From mboxrd@z Thu Jan 1 00:00:00 1970 Return-Path: Received: (qmail 106366 invoked by alias); 28 Apr 2015 18:46:06 -0000 Mailing-List: contact gcc-patches-help@gcc.gnu.org; run by ezmlm Precedence: bulk List-Id: List-Archive: List-Post: List-Help: Sender: gcc-patches-owner@gcc.gnu.org Received: (qmail 106355 invoked by uid 89); 28 Apr 2015 18:46:05 -0000 Authentication-Results: sourceware.org; auth=none X-Virus-Found: No X-Spam-SWARE-Status: No, score=-2.3 required=5.0 tests=AWL,BAYES_00,RCVD_IN_DNSWL_LOW,SPF_PASS autolearn=ham version=3.3.2 X-HELO: relay1.mentorg.com Received: from relay1.mentorg.com (HELO relay1.mentorg.com) (192.94.38.131) by sourceware.org (qpsmtpd/0.93/v0.84-503-g423c35a) with ESMTP; Tue, 28 Apr 2015 18:46:04 +0000 Received: from nat-ies.mentorg.com ([192.94.31.2] helo=SVR-IES-FEM-02.mgc.mentorg.com) by relay1.mentorg.com with esmtp id 1YnAW7-0006Kb-KN from Thomas_Schwinge@mentor.com ; Tue, 28 Apr 2015 11:45:59 -0700 Received: from feldtkeller.schwinge.homeip.net (137.202.0.76) by SVR-IES-FEM-02.mgc.mentorg.com (137.202.0.106) with Microsoft SMTP Server id 14.3.224.2; Tue, 28 Apr 2015 19:45:58 +0100 From: Thomas Schwinge To: , Jakub Jelinek CC: Cesar Philippidis Subject: Fix OpenMP's target update directive in templated code In-Reply-To: <553E787A.1020109@mentor.com> References: <553E695A.2070007@mentor.com> <87zj5ttqpz.fsf@schwinge.name> <553E787A.1020109@mentor.com> User-Agent: Notmuch/0.9-101-g81dad07 (http://notmuchmail.org) Emacs/24.3.1 (x86_64-pc-linux-gnu) Date: Tue, 28 Apr 2015 19:30:00 -0000 Message-ID: <87383kt7kx.fsf@schwinge.name> MIME-Version: 1.0 Content-Type: multipart/signed; boundary="=-=-="; micalg=pgp-sha1; protocol="application/pgp-signature" X-SW-Source: 2015-04/txt/msg01788.txt.bz2 --=-=-= Content-Type: text/plain; charset=utf-8 Content-Transfer-Encoding: quoted-printable Content-length: 6322 Hi Jakub! When adding support for C++ templates usage with OpenACC directives (gcc/cp/pt.c:tsubst_expr), we found: On Mon, 27 Apr 2015 10:57:14 -0700, Cesar Philippidis wrote: > On 04/27/2015 10:40 AM, Thomas Schwinge wrote: > > Cesar wrote: > >> case OMP_TARGET_UPDATE: > >> tmp =3D tsubst_omp_clauses (OMP_TARGET_UPDATE_CLAUSES (t), fals= e, > >> args, complain, in_decl); > >> @@ -14253,6 +14292,16 @@ tsubst_expr (tree t, tree args, tsubst_flags_= t complain, tree in_decl, > >> add_stmt (t); > >> break; > >>=20=20 > >> + case OACC_ENTER_DATA: > >> + case OACC_EXIT_DATA: > >> + case OACC_UPDATE: > >> + tmp =3D tsubst_omp_clauses (TREE_OPERAND (t, 0), false, > >> + args, complain, in_decl); > >> + t =3D copy_node (t); > >> + TREE_OPERAND (t, 0) =3D tmp; > >> + add_stmt (t); > >> + break; > >=20 > > Given the handling for similar codes, why not handle those together with > > OMP_TARGET_UPDATE, replacing OMP_TARGET_UPDATE_CLAUSES with plain > > OMP_CLAUSES? >=20 > I left it them separate because OMP_CLAUSES expects the clauses to > appear in tree_operand 1, but oacc_update, oacc_enter_data and > oacc_exit_data have the clauses at tree_operand 0. I think this may have > been a bug with openmp, so I was planning on informing Jakub when I > posted this patch upstream. >=20 > The attached patch cleans up [...] > --- a/gcc/cp/pt.c > +++ b/gcc/cp/pt.c > @@ -14277,21 +14277,7 @@ tsubst_expr (tree t, tree args, tsubst_flags_t c= omplain, tree in_decl, > add_stmt (t); > break; >=20=20 > case OMP_TARGET_UPDATE: > - tmp =3D tsubst_omp_clauses (OMP_TARGET_UPDATE_CLAUSES (t), false, > - args, complain, in_decl); > - t =3D copy_node (t); > - OMP_CLAUSES (t) =3D tmp; > - add_stmt (t); > - break; > - > case OACC_ENTER_DATA: > case OACC_EXIT_DATA: > case OACC_UPDATE: I guess nobody so far ;-) has been using OpenMP's target update directive in templated code -- OK to commit the following, and to which branches (4.9, 5, trunk)? commit 5ea85b95c1d0ccb98d73345f105bf76839b16433 Author: Thomas Schwinge Date: Tue Apr 28 20:29:26 2015 +0200 Fix OpenMP's target update directive in templated code. =20=20=20=20 FAIL: g++.dg/gomp/tpl-target-update.C -std=3Dc++98 (internal compi= ler error) FAIL: g++.dg/gomp/tpl-target-update.C -std=3Dc++98 (test for exces= s errors) FAIL: g++.dg/gomp/tpl-target-update.C -std=3Dc++11 (internal compi= ler error) FAIL: g++.dg/gomp/tpl-target-update.C -std=3Dc++11 (test for exces= s errors) FAIL: g++.dg/gomp/tpl-target-update.C -std=3Dc++14 (internal compi= ler error) FAIL: g++.dg/gomp/tpl-target-update.C -std=3Dc++14 (test for exces= s errors) =20=20=20=20 [...]/source-gcc/gcc/testsuite/g++.dg/gomp/tpl-target-update.C: In = instantiation of 'void f(T, T) [with T =3D int]': [...]/source-gcc/gcc/testsuite/g++.dg/gomp/tpl-target-update.C:19:9= : required from here [...]/source-gcc/gcc/testsuite/g++.dg/gomp/tpl-target-update.C:10:9= : internal compiler error: tree check: expected oacc_parallel or oacc_kerne= ls or oacc_data or oacc_host_data or omp_parallel or omp_task or omp_for or= omp_simd or cilk_simd or cilk_for or omp_distribute or oacc_loop or omp_te= ams or omp_target_data or omp_target or omp_sections or omp_single, have om= p_target_update in tsubst_expr, at cp/pt.c:14209 0xf5aae1 tree_range_check_failed(tree_node const*, char const*, int= , char const*, tree_code, tree_code) [...]/source-gcc/gcc/tree.c:9384 0x66e201 tree_range_check [...]/source-gcc/gcc/tree.h:2979 0x66e201 tsubst_expr [...]/source-gcc/gcc/cp/pt.c:14209 0x6695e3 tsubst_expr [...]/source-gcc/gcc/cp/pt.c:13752 0x66ac07 tsubst_expr [...]/source-gcc/gcc/cp/pt.c:13938 0x667c41 instantiate_decl(tree_node*, int, bool) [...]/source-gcc/gcc/cp/pt.c:20367 0x6ae386 instantiate_pending_templates(int) [...]/source-gcc/gcc/cp/pt.c:20484 0x6edc3d cp_write_global_declarations() [...]/source-gcc/gcc/cp/decl2.c:4456 =20=20=20=20 gcc/cp/ * pt.c (tsubst_expr) : Use OMP_TARGET_UPDATE_CLAUSES instead of OMP_CLAUSES. gcc/testsuite/ * g++.dg/gomp/tpl-target-update.C: New file. --- gcc/cp/pt.c | 2 +- gcc/testsuite/g++.dg/gomp/tpl-target-update.C | 20 ++++++++++++++++++++ 2 files changed, 21 insertions(+), 1 deletion(-) diff --git gcc/cp/pt.c gcc/cp/pt.c index f9a5c3b..129517a 100644 --- gcc/cp/pt.c +++ gcc/cp/pt.c @@ -14206,7 +14206,7 @@ tsubst_expr (tree t, tree args, tsubst_flags_t comp= lain, tree in_decl, tmp =3D tsubst_omp_clauses (OMP_TARGET_UPDATE_CLAUSES (t), false, args, complain, in_decl); t =3D copy_node (t); - OMP_CLAUSES (t) =3D tmp; + OMP_TARGET_UPDATE_CLAUSES (t) =3D tmp; add_stmt (t); break; =20 diff --git gcc/testsuite/g++.dg/gomp/tpl-target-update.C gcc/testsuite/g++.= dg/gomp/tpl-target-update.C new file mode 100644 index 0000000..6226ebf --- /dev/null +++ gcc/testsuite/g++.dg/gomp/tpl-target-update.C @@ -0,0 +1,20 @@ +// { dg-do compile } + +template +void f(T A, T B) +{ + extern int *v; + T a =3D 2; + T b =3D 4; + +#pragma omp target update to(v[a:b]) + v[a] =3D 0; + +#pragma omp target update to(v[A:B]) + v[a] =3D 0; +} + +void g() +{ + f(1, 5); +} That said, what is the preferred approach to add support for OACC_ENTER_DATA, OACC_EXIT_DATA, OACC_UPDATE? I'm not sure hard-coding TREE_OPERAND (t, 0) in gcc/cp/pt.c:tsubst_expr is the way to go, and also duplicating the OMP_TARGET_UPDATE code for each of the three OACC_* doesn't sound appealing -- especially given that we just "switch"ed on the respective tree code, so the O*_CHECK of the respective O*_CLAUSES macro is completely redundant. Is it OK to extend gcc/tree.h:OMP_CLAUSES so that it can also be used for tree codes that keep clauses in operand 0, and then use that here (and also in gcc/gimplify.c:gimplify_omp_target_update, for example)? Gr=C3=BC=C3=9Fe, Thomas --=-=-= Content-Type: application/pgp-signature Content-length: 472 -----BEGIN PGP SIGNATURE----- Version: GnuPG v1 iQEcBAEBAgAGBQJVP9VeAAoJEPoxNhtoi6CON3EH/2N+VXlseqB5t/OJGNDL2joA Kre+b31n0/idnQhO+ZwXbufAGbxoy2JNhl3+MtjKHfn5w7+hkMsd4vAzxbmuvQQh Vs5irqufSuDofIokmgRAvI2Iq7LgqbTZoHP8vLN8L5wsShAKYm+u2aajnfTB6m68 k1SybDAkaaTsm283q5P/Ru4kO6E//KrJJ57YkAyiLbQIzFyU82YZKALf3aR1bzUm 88WHCo4pygkCUlCUbv/0oxPfykUcYDNFVk4XHi3usjrkBmAtJRN/a012mxeFe/Ia IkO7h9fGDvgqrdl4Cuh8Bt8ofRZRZslhEszNGuMzNpE1/TeQ786+LOO6NY23Yv4= =vzik -----END PGP SIGNATURE----- --=-=-=--