Fix OpenMP's target update directive in templated code

Fix OpenMP's target update directive in templated code

[Thomas Schwinge]

[-- Attachment #1: Type: text/plain, Size: 6347 bytes --]

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 <cesar_philippidis@mentor.com> wrote:
> On 04/27/2015 10:40 AM, Thomas Schwinge wrote:
> > Cesar wrote:
> >>      case OMP_TARGET_UPDATE:
> >>        tmp = tsubst_omp_clauses (OMP_TARGET_UPDATE_CLAUSES (t), false,
> >>  				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;
> >>  
> >> +    case OACC_ENTER_DATA:
> >> +    case OACC_EXIT_DATA:
> >> +    case OACC_UPDATE:
> >> +      tmp = tsubst_omp_clauses (TREE_OPERAND (t, 0), false,
> >> +				args, complain, in_decl);
> >> +      t = copy_node (t);
> >> +      TREE_OPERAND (t, 0) = tmp;
> >> +      add_stmt (t);
> >> +      break;
> > 
> > Given the handling for similar codes, why not handle those together with
> > OMP_TARGET_UPDATE, replacing OMP_TARGET_UPDATE_CLAUSES with plain
> > OMP_CLAUSES?
> 
> 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.
> 
> 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 complain, tree in_decl,
>        add_stmt (t);
>        break;
>  
>      case OMP_TARGET_UPDATE:
> -      tmp = tsubst_omp_clauses (OMP_TARGET_UPDATE_CLAUSES (t), false,
> -				args, complain, in_decl);
> -      t = copy_node (t);
> -      OMP_CLAUSES (t) = 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 <thomas@codesourcery.com>
Date:   Tue Apr 28 20:29:26 2015 +0200

    Fix OpenMP's target update directive in templated code.
    
        FAIL: g++.dg/gomp/tpl-target-update.C  -std=c++98 (internal compiler error)
        FAIL: g++.dg/gomp/tpl-target-update.C  -std=c++98 (test for excess errors)
        FAIL: g++.dg/gomp/tpl-target-update.C  -std=c++11 (internal compiler error)
        FAIL: g++.dg/gomp/tpl-target-update.C  -std=c++11 (test for excess errors)
        FAIL: g++.dg/gomp/tpl-target-update.C  -std=c++14 (internal compiler error)
        FAIL: g++.dg/gomp/tpl-target-update.C  -std=c++14 (test for excess errors)
    
        [...]/source-gcc/gcc/testsuite/g++.dg/gomp/tpl-target-update.C: In instantiation of 'void f(T, T) [with T = 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_kernels 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_teams or omp_target_data or omp_target or omp_sections or omp_single, have omp_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
    
    	gcc/cp/
    	* pt.c (tsubst_expr) <OMP_TARGET_UPDATE>: 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 complain, tree in_decl,
       tmp = tsubst_omp_clauses (OMP_TARGET_UPDATE_CLAUSES (t), false,
 				args, complain, in_decl);
       t = copy_node (t);
-      OMP_CLAUSES (t) = tmp;
+      OMP_TARGET_UPDATE_CLAUSES (t) = tmp;
       add_stmt (t);
       break;
 
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 <typename T>
+void f(T A, T B)
+{
+  extern int *v;
+  T a = 2;
+  T b = 4;
+
+#pragma omp target update to(v[a:b])
+  v[a] = 0;
+
+#pragma omp target update to(v[A:B])
+  v[a] = 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üße,
 Thomas

[-- Attachment #2: Type: application/pgp-signature, Size: 472 bytes --]

Re: Fix OpenMP's target update directive in templated code

[Jakub Jelinek]

On Tue, Apr 28, 2015 at 08:45:50PM +0200, Thomas Schwinge wrote:
> 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)?

Seems I've missed testcases for target {,update,data} in templates indeed,
generally for C++ I'm trying to add testcases for templates, both when
relevant types are type dependent and whey aren't.

>     	gcc/cp/
>     	* pt.c (tsubst_expr) <OMP_TARGET_UPDATE>: Use
>     	OMP_TARGET_UPDATE_CLAUSES instead of OMP_CLAUSES.
>     	gcc/testsuite/
>     	* g++.dg/gomp/tpl-target-update.C: New file.

This is ok for trunk, 5.2 and 4.9.3, thanks for fixing that.

> 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)?

How could it work when it is operand 1 on some and operand 0 in others.
IMHO, if you want to reuse the same code for OMP_TARGET_UPDATE,
various OACC_* standalone directives (and
OMP_TARGET_ENTER_DATA/OMP_TARGET_EXIT_DATA in OpenMP 4.1), then
you should make sure they are consecutive in target.def and
define OMP_STANDALONE_CLAUSES or OMP_TARGET_STANDALONE_CLAUSES as
a range check between OMP_TARGET_UPDATE and the last OpenACC directive
without body, just clauses.

	Jakub

Re: Fix OpenMP's target update directive in templated code

[Thomas Schwinge]

[-- Attachment #1: Type: text/plain, Size: 911 bytes --]

Hi Jakub!

On Wed, 29 Apr 2015 10:53:32 +0200, Jakub Jelinek <jakub@redhat.com> wrote:
> On Tue, Apr 28, 2015 at 08:45:50PM +0200, Thomas Schwinge wrote:
> > 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)?
> 
> Seems I've missed testcases for target {,update,data} in templates indeed,
> generally for C++ I'm trying to add testcases for templates, both when
> relevant types are type dependent and whey aren't.
> 
> >     	gcc/cp/
> >     	* pt.c (tsubst_expr) <OMP_TARGET_UPDATE>: Use
> >     	OMP_TARGET_UPDATE_CLAUSES instead of OMP_CLAUSES.
> >     	gcc/testsuite/
> >     	* g++.dg/gomp/tpl-target-update.C: New file.
> 
> This is ok for trunk, 5.2 and 4.9.3, thanks for fixing that.

Committed in r222564, r222565, r222566, respectively.


Grüße,
 Thomas

[-- Attachment #2: Type: application/pgp-signature, Size: 472 bytes --]

OMP_CLAUSES with clauses in operand 0 (was: Fix OpenMP's target update directive in templated code)

[Thomas Schwinge]

[-- Attachment #1: Type: text/plain, Size: 2225 bytes --]

Hi Jakub!

On Wed, 29 Apr 2015 10:53:32 +0200, Jakub Jelinek <jakub@redhat.com> wrote:
> On Tue, Apr 28, 2015 at 08:45:50PM +0200, Thomas Schwinge wrote:
> > 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)?
> 
> How could it work when it is operand 1 on some and operand 0 in others.

Something like (untested):

     #define OMP_CLAUSES(NODE) \
    -  TREE_OPERAND (TREE_RANGE_CHECK (NODE, OACC_PARALLEL, OMP_SINGLE), 1) \
    +  (tree_code (NODE) <= OMP_SINGLE) \
    +  ? TREE_OPERAND (TREE_RANGE_CHECK (NODE, OACC_PARALLEL, OMP_SINGLE), 1) \
    +  : TREE_OPERAND (TREE_RANGE_CHECK (NODE, OACC_CACHE, OMP_TARGET_UPDATE), 0)

Rationale: I'm not expecting another variant to be added later on
(clauses are either in operand 0 or 1).  Encoding explicit tree code
(ordering) assuptions is already done with the usage of TREE_RANGE_CHECK
macros, so the additional tree code check doesn't make this much worse.
It has the benefit of offering the same known OMP_CLAUSES interface.

Yet, if that's a non-starter, I'll pursue this one:

> IMHO, if you want to reuse the same code for OMP_TARGET_UPDATE,
> various OACC_* standalone directives (and
> OMP_TARGET_ENTER_DATA/OMP_TARGET_EXIT_DATA in OpenMP 4.1), then
> you should make sure they are consecutive in target.def

(tree.def; we've already made sure they're grouped consecutively.)

> and
> define OMP_STANDALONE_CLAUSES or OMP_TARGET_STANDALONE_CLAUSES as
> a range check between OMP_TARGET_UPDATE and the last OpenACC directive
> without body, just clauses.


Grüße,
 Thomas

[-- Attachment #2: Type: application/pgp-signature, Size: 472 bytes --]

Re: OMP_CLAUSES with clauses in operand 0 (was: Fix OpenMP's target update directive in templated code)

[Jakub Jelinek]

On Wed, Apr 29, 2015 at 11:28:55AM +0200, Thomas Schwinge wrote:
> Yet, if that's a non-starter, I'll pursue this one:

Yeah, it is a non-starter, it has unnecessary runtime overhead everywhere
where it is used.

	Jakub

Re: OMP_CLAUSES with clauses in operand 0

[Thomas Schwinge]

[-- Attachment #1: Type: text/plain, Size: 2508 bytes --]

Hi!

On Wed, 29 Apr 2015 11:32:31 +0200, Jakub Jelinek <jakub@redhat.com> wrote:
> Yeah, it is a non-starter, it has unnecessary runtime overhead everywhere
> where it is used.

Huh.  OMP_CLAUSES is currently used in a dozen places in
cp/cp-gimplify.c, cp/pt.c, and gimplify.c.  I've been expecting my
changed code to be well-optimizable, for the compiler already knows
TREE_CODE(NODE) in a lot of these places.  Doing a quick before (1)/after
(2) comparison test with -g -O2 on x86_64 GNU/Linux using GCC 4.8.3, the
object file sizes are as follows:

       text    data     bss     dec     hex filename
      37027       0      16   37043    90b3 1/build-gcc/gcc/cp/cp-gimplify.o
      36307       0      16   36323    8de3 2/build-gcc/gcc/cp/cp-gimplify.o
       text    data     bss     dec     hex filename
     458742       0     136  458878   7007e 1/build-gcc/gcc/cp/pt.o
     458630       0     136  458766   7000e 2/build-gcc/gcc/cp/pt.o
       text    data     bss     dec     hex filename
     166056       0      48  166104   288d8 1/build-gcc/gcc/gimplify.o
     166200       0      48  166248   28968 2/build-gcc/gcc/gimplify.o

..., so actually smaller for the first two files.  Admittedly, there is a
144 bytes (0.0867 %) size increase in gimplify.o, and I have not measured
the actual runtime overhead (which I'm totally expecting to be "in the
noise", but...), so I'm assuming that my proposal to keep the interface
simple does not pay for the presumed runtime overhead, so I guess I'm
posting this patch just for the archives...

--- gcc/tree.h
+++ gcc/tree.h
@@ -1200,7 +1200,9 @@ extern void protected_set_expr_location (tree, location_t);
 #define OMP_BODY(NODE) \
   TREE_OPERAND (TREE_RANGE_CHECK (NODE, OACC_PARALLEL, OMP_CRITICAL), 0)
 #define OMP_CLAUSES(NODE) \
-  TREE_OPERAND (TREE_RANGE_CHECK (NODE, OACC_PARALLEL, OMP_SINGLE), 1)
+  ((TREE_CODE (NODE) <= OMP_SINGLE) \
+   ? TREE_OPERAND (TREE_RANGE_CHECK (NODE, OACC_PARALLEL, OMP_SINGLE), 1) \
+   : TREE_OPERAND (TREE_RANGE_CHECK (NODE, OACC_CACHE, OMP_TARGET_UPDATE), 0))
 
 #define OACC_PARALLEL_BODY(NODE) \
   TREE_OPERAND (OACC_PARALLEL_CHECK (NODE), 0)

If that's not been obvious, I favor code readability (one generic
OMP_CLAUSES interface instead of having both OMP_CLAUSES and
OMP_CLAUSES_STANDALONE) over a tiny code size regression, but you seem to
think differently, set a different policy for code changes, which I'll
have to yield to.


Grüße,
 Thomas

[-- Attachment #2: Type: application/pgp-signature, Size: 472 bytes --]

Re: OMP_CLAUSES with clauses in operand 0

[Jakub Jelinek]

On Wed, Apr 29, 2015 at 01:13:29PM +0200, Thomas Schwinge wrote:
> On Wed, 29 Apr 2015 11:32:31 +0200, Jakub Jelinek <jakub@redhat.com> wrote:
> > Yeah, it is a non-starter, it has unnecessary runtime overhead everywhere
> > where it is used.
> 
> Huh.  OMP_CLAUSES is currently used in a dozen places in
> cp/cp-gimplify.c, cp/pt.c, and gimplify.c.  I've been expecting my
> changed code to be well-optimizable, for the compiler already knows
> TREE_CODE(NODE) in a lot of these places.  Doing a quick before (1)/after
> (2) comparison test with -g -O2 on x86_64 GNU/Linux using GCC 4.8.3, the
> object file sizes are as follows:
> 
>        text    data     bss     dec     hex filename
>       37027       0      16   37043    90b3 1/build-gcc/gcc/cp/cp-gimplify.o
>       36307       0      16   36323    8de3 2/build-gcc/gcc/cp/cp-gimplify.o
>        text    data     bss     dec     hex filename
>      458742       0     136  458878   7007e 1/build-gcc/gcc/cp/pt.o
>      458630       0     136  458766   7000e 2/build-gcc/gcc/cp/pt.o
>        text    data     bss     dec     hex filename
>      166056       0      48  166104   288d8 1/build-gcc/gcc/gimplify.o
>      166200       0      48  166248   28968 2/build-gcc/gcc/gimplify.o
> 
> ..., so actually smaller for the first two files.  Admittedly, there is a
> 144 bytes (0.0867 %) size increase in gimplify.o, and I have not measured
> the actual runtime overhead (which I'm totally expecting to be "in the
> noise", but...), so I'm assuming that my proposal to keep the interface
> simple does not pay for the presumed runtime overhead, so I guess I'm
> posting this patch just for the archives...

I really don't understand how you could get smaller code out of that, that
doesn't make any sense.  And I don't see where it would make sense to share
the same code for handling the standalone directives and directives with
associated blocks, in all the places I've looked at you want a different
code.  And in many cases you don't have just a single TREE_CODE, but
multiple, likely all of them from the same category (either all of them
smaller/equal or all larger than OMP_SINGLE), but VRP doesn't have to be
able to fix up the weirdnesses.

So yes, I really prefer OMP_STANDALONE_CLAUSES over OMP_CLAUSES for
everything.  And, tree checking which is on by default should catch this
properly, it was just a matter of tests not being written when they should
have been.

	Jakub

Re: OMP_CLAUSES with clauses in operand 0

[Thomas Schwinge]

[-- Attachment #1: Type: text/plain, Size: 13966 bytes --]

Hi Jakub!

On Wed, 29 Apr 2015 13:43:55 +0200, Jakub Jelinek <jakub@redhat.com> wrote:
> On Wed, Apr 29, 2015 at 01:13:29PM +0200, Thomas Schwinge wrote:
> > On Wed, 29 Apr 2015 11:32:31 +0200, Jakub Jelinek <jakub@redhat.com> wrote:
> > > Yeah, it is a non-starter, it has unnecessary runtime overhead everywhere
> > > where it is used.
> > 
> > Huh.  OMP_CLAUSES is currently used in a dozen places in
> > cp/cp-gimplify.c, cp/pt.c, and gimplify.c.  I've been expecting my
> > changed code to be well-optimizable, for the compiler already knows
> > TREE_CODE(NODE) in a lot of these places.  Doing a quick before (1)/after
> > (2) comparison test with -g -O2 on x86_64 GNU/Linux using GCC 4.8.3, the
> > object file sizes are as follows:
> > 
> >        text    data     bss     dec     hex filename
> >       37027       0      16   37043    90b3 1/build-gcc/gcc/cp/cp-gimplify.o
> >       36307       0      16   36323    8de3 2/build-gcc/gcc/cp/cp-gimplify.o
> >        text    data     bss     dec     hex filename
> >      458742       0     136  458878   7007e 1/build-gcc/gcc/cp/pt.o
> >      458630       0     136  458766   7000e 2/build-gcc/gcc/cp/pt.o
> >        text    data     bss     dec     hex filename
> >      166056       0      48  166104   288d8 1/build-gcc/gcc/gimplify.o
> >      166200       0      48  166248   28968 2/build-gcc/gcc/gimplify.o
> > 
> > ..., so actually smaller for the first two files.  Admittedly, there is a
> > 144 bytes (0.0867 %) size increase in gimplify.o, and I have not measured
> > the actual runtime overhead (which I'm totally expecting to be "in the
> > noise", but...), so I'm assuming that my proposal to keep the interface
> > simple does not pay for the presumed runtime overhead, so I guess I'm
> > posting this patch just for the archives...
> 
> I really don't understand how you could get smaller code out of that, that
> doesn't make any sense.

I tried a quick -fdump-tree-all comparison but that doesn't lead
anywhere, because a vast number of IDs change.  Any tricks how to tackle
such a thing?


> So yes, I really prefer OMP_STANDALONE_CLAUSES over OMP_CLAUSES for
> everything.

Like this (for trunk)?

commit 300e28fce192cb56d73cb61f787872643030f0bf
Author: Thomas Schwinge <thomas@codesourcery.com>
Date:   Wed Apr 29 16:18:49 2015 +0200

    Add OMP_STANDALONE_CLAUSES.
    
    	gcc/
    	* tree.h (OACC_CACHE_CLAUSES, OACC_DECLARE_CLAUSES)
    	(OACC_ENTER_DATA_CLAUSES, OACC_EXIT_DATA_CLAUSES)
    	(OACC_UPDATE_CLAUSES, OMP_TARGET_UPDATE_CLAUSES): Merge into...
    	(OMP_STANDALONE_CLAUSES): ... this new macro.  Adjust all users.
---
 gcc/c/c-parser.c        |   11 ++++-------
 gcc/cp/parser.c         |   11 ++++-------
 gcc/cp/pt.c             |    4 ++--
 gcc/gimplify.c          |   18 ++++++++----------
 gcc/tree-pretty-print.c |   12 ++++++------
 gcc/tree.def            |   12 ++++++------
 gcc/tree.h              |   19 ++-----------------
 7 files changed, 32 insertions(+), 55 deletions(-)

diff --git gcc/c/c-parser.c gcc/c/c-parser.c
index f5e2ac2c..86b77f3 100644
--- gcc/c/c-parser.c
+++ gcc/c/c-parser.c
@@ -11987,7 +11987,7 @@ c_parser_oacc_cache (location_t loc, c_parser *parser)
 
   stmt = make_node (OACC_CACHE);
   TREE_TYPE (stmt) = void_type_node;
-  OACC_CACHE_CLAUSES (stmt) = clauses;
+  OMP_STANDALONE_CLAUSES (stmt) = clauses;
   SET_EXPR_LOCATION (stmt, loc);
   add_stmt (stmt);
 
@@ -12155,10 +12155,7 @@ c_parser_oacc_enter_exit_data (c_parser *parser, bool enter)
 
   stmt = enter ? make_node (OACC_ENTER_DATA) : make_node (OACC_EXIT_DATA);
   TREE_TYPE (stmt) = void_type_node;
-  if (enter)
-    OACC_ENTER_DATA_CLAUSES (stmt) = clauses;
-  else
-    OACC_EXIT_DATA_CLAUSES (stmt) = clauses;
+  OMP_STANDALONE_CLAUSES (stmt) = clauses;
   SET_EXPR_LOCATION (stmt, loc);
   add_stmt (stmt);
 }
@@ -12285,7 +12282,7 @@ c_parser_oacc_update (c_parser *parser)
 
   tree stmt = make_node (OACC_UPDATE);
   TREE_TYPE (stmt) = void_type_node;
-  OACC_UPDATE_CLAUSES (stmt) = clauses;
+  OMP_STANDALONE_CLAUSES (stmt) = clauses;
   SET_EXPR_LOCATION (stmt, loc);
   add_stmt (stmt);
 }
@@ -13858,7 +13855,7 @@ c_parser_omp_target_update (location_t loc, c_parser *parser,
 
   tree stmt = make_node (OMP_TARGET_UPDATE);
   TREE_TYPE (stmt) = void_type_node;
-  OMP_TARGET_UPDATE_CLAUSES (stmt) = clauses;
+  OMP_STANDALONE_CLAUSES (stmt) = clauses;
   SET_EXPR_LOCATION (stmt, loc);
   add_stmt (stmt);
   return false;
diff --git gcc/cp/parser.c gcc/cp/parser.c
index 4ea2ca2..61fd34f 100644
--- gcc/cp/parser.c
+++ gcc/cp/parser.c
@@ -31386,7 +31386,7 @@ cp_parser_omp_target_update (cp_parser *parser, cp_token *pragma_tok,
 
   tree stmt = make_node (OMP_TARGET_UPDATE);
   TREE_TYPE (stmt) = void_type_node;
-  OMP_TARGET_UPDATE_CLAUSES (stmt) = clauses;
+  OMP_STANDALONE_CLAUSES (stmt) = clauses;
   SET_EXPR_LOCATION (stmt, pragma_tok->location);
   add_stmt (stmt);
   return false;
@@ -31496,7 +31496,7 @@ cp_parser_oacc_cache (cp_parser *parser, cp_token *pragma_tok)
 
   stmt = make_node (OACC_CACHE);
   TREE_TYPE (stmt) = void_type_node;
-  OACC_CACHE_CLAUSES (stmt) = clauses;
+  OMP_STANDALONE_CLAUSES (stmt) = clauses;
   SET_EXPR_LOCATION (stmt, pragma_tok->location);
   add_stmt (stmt);
 
@@ -31606,10 +31606,7 @@ cp_parser_oacc_enter_exit_data (cp_parser *parser, cp_token *pragma_tok,
 
   stmt = enter ? make_node (OACC_ENTER_DATA) : make_node (OACC_EXIT_DATA);
   TREE_TYPE (stmt) = void_type_node;
-  if (enter)
-    OACC_ENTER_DATA_CLAUSES (stmt) = clauses;
-  else
-    OACC_EXIT_DATA_CLAUSES (stmt) = clauses;
+  OMP_STANDALONE_CLAUSES (stmt) = clauses;
   SET_EXPR_LOCATION (stmt, pragma_tok->location);
   add_stmt (stmt);
   return stmt;
@@ -31746,7 +31743,7 @@ cp_parser_oacc_update (cp_parser *parser, cp_token *pragma_tok)
 
   stmt = make_node (OACC_UPDATE);
   TREE_TYPE (stmt) = void_type_node;
-  OACC_UPDATE_CLAUSES (stmt) = clauses;
+  OMP_STANDALONE_CLAUSES (stmt) = clauses;
   SET_EXPR_LOCATION (stmt, pragma_tok->location);
   add_stmt (stmt);
   return stmt;
diff --git gcc/cp/pt.c gcc/cp/pt.c
index 129517a..9e83c22 100644
--- gcc/cp/pt.c
+++ gcc/cp/pt.c
@@ -14203,10 +14203,10 @@ tsubst_expr (tree t, tree args, tsubst_flags_t complain, tree in_decl,
       break;
 
     case OMP_TARGET_UPDATE:
-      tmp = tsubst_omp_clauses (OMP_TARGET_UPDATE_CLAUSES (t), false,
+      tmp = tsubst_omp_clauses (OMP_STANDALONE_CLAUSES (t), false,
 				args, complain, in_decl);
       t = copy_node (t);
-      OMP_TARGET_UPDATE_CLAUSES (t) = tmp;
+      OMP_STANDALONE_CLAUSES (t) = tmp;
       add_stmt (t);
       break;
 
diff --git gcc/gimplify.c gcc/gimplify.c
index c68bd47..a68748a 100644
--- gcc/gimplify.c
+++ gcc/gimplify.c
@@ -6799,8 +6799,9 @@ gimplify_oacc_cache (tree *expr_p, gimple_seq *pre_p)
 {
   tree expr = *expr_p;
 
-  gimplify_scan_omp_clauses (&OACC_CACHE_CLAUSES (expr), pre_p, ORT_WORKSHARE);
-  gimplify_adjust_omp_clauses (pre_p, &OACC_CACHE_CLAUSES (expr));
+  gimplify_scan_omp_clauses (&OMP_STANDALONE_CLAUSES (expr), pre_p,
+			     ORT_WORKSHARE);
+  gimplify_adjust_omp_clauses (pre_p, &OMP_STANDALONE_CLAUSES (expr));
 
   /* TODO: Do something sensible with this information.  */
 
@@ -7427,34 +7428,31 @@ gimplify_omp_workshare (tree *expr_p, gimple_seq *pre_p)
 static void
 gimplify_omp_target_update (tree *expr_p, gimple_seq *pre_p)
 {
-  tree expr = *expr_p, clauses;
+  tree expr = *expr_p;
   int kind;
   gomp_target *stmt;
 
   switch (TREE_CODE (expr))
     {
     case OACC_ENTER_DATA:
-      clauses = OACC_ENTER_DATA_CLAUSES (expr);
       kind = GF_OMP_TARGET_KIND_OACC_ENTER_EXIT_DATA;
       break;
     case OACC_EXIT_DATA:
-      clauses = OACC_EXIT_DATA_CLAUSES (expr);
       kind = GF_OMP_TARGET_KIND_OACC_ENTER_EXIT_DATA;
       break;
     case OACC_UPDATE:
-      clauses = OACC_UPDATE_CLAUSES (expr);
       kind = GF_OMP_TARGET_KIND_OACC_UPDATE;
       break;
     case OMP_TARGET_UPDATE:
-      clauses = OMP_TARGET_UPDATE_CLAUSES (expr);
       kind = GF_OMP_TARGET_KIND_UPDATE;
       break;
     default:
       gcc_unreachable ();
     }
-  gimplify_scan_omp_clauses (&clauses, pre_p, ORT_WORKSHARE);
-  gimplify_adjust_omp_clauses (pre_p, &clauses);
-  stmt = gimple_build_omp_target (NULL, kind, clauses);
+  gimplify_scan_omp_clauses (&OMP_STANDALONE_CLAUSES (expr), pre_p,
+			     ORT_WORKSHARE);
+  gimplify_adjust_omp_clauses (pre_p, &OMP_STANDALONE_CLAUSES (expr));
+  stmt = gimple_build_omp_target (NULL, kind, OMP_STANDALONE_CLAUSES (expr));
 
   gimplify_seq_add_stmt (pre_p, stmt);
   *expr_p = NULL_TREE;
diff --git gcc/tree-pretty-print.c gcc/tree-pretty-print.c
index d7c049f..26f517f 100644
--- gcc/tree-pretty-print.c
+++ gcc/tree-pretty-print.c
@@ -2585,27 +2585,27 @@ dump_generic_node (pretty_printer *pp, tree node, int spc, int flags,
 
     case OACC_DECLARE:
       pp_string (pp, "#pragma acc declare");
-      dump_omp_clauses (pp, OACC_DECLARE_CLAUSES (node), spc, flags);
+      dump_omp_clauses (pp, OMP_STANDALONE_CLAUSES (node), spc, flags);
       break;
 
     case OACC_UPDATE:
       pp_string (pp, "#pragma acc update");
-      dump_omp_clauses (pp, OACC_UPDATE_CLAUSES (node), spc, flags);
+      dump_omp_clauses (pp, OMP_STANDALONE_CLAUSES (node), spc, flags);
       break;
 
     case OACC_ENTER_DATA:
       pp_string (pp, "#pragma acc enter data");
-      dump_omp_clauses (pp, OACC_ENTER_DATA_CLAUSES (node), spc, flags);
+      dump_omp_clauses (pp, OMP_STANDALONE_CLAUSES (node), spc, flags);
       break;
 
     case OACC_EXIT_DATA:
       pp_string (pp, "#pragma acc exit data");
-      dump_omp_clauses (pp, OACC_EXIT_DATA_CLAUSES (node), spc, flags);
+      dump_omp_clauses (pp, OMP_STANDALONE_CLAUSES (node), spc, flags);
       break;
 
     case OACC_CACHE:
       pp_string (pp, "#pragma acc cache");
-      dump_omp_clauses (pp, OACC_CACHE_CLAUSES (node), spc, flags);
+      dump_omp_clauses (pp, OMP_STANDALONE_CLAUSES (node), spc, flags);
       break;
 
     case OMP_PARALLEL:
@@ -2673,7 +2673,7 @@ dump_generic_node (pretty_printer *pp, tree node, int spc, int flags,
 
     case OMP_TARGET_UPDATE:
       pp_string (pp, "#pragma omp target update");
-      dump_omp_clauses (pp, OMP_TARGET_UPDATE_CLAUSES (node), spc, flags);
+      dump_omp_clauses (pp, OMP_STANDALONE_CLAUSES (node), spc, flags);
       is_expr = false;
       break;
 
diff --git gcc/tree.def gcc/tree.def
index b4b4164..2dd70b9 100644
--- gcc/tree.def
+++ gcc/tree.def
@@ -1157,28 +1157,28 @@ DEFTREECODE (OMP_ORDERED, "omp_ordered", tcc_statement, 1)
 DEFTREECODE (OMP_CRITICAL, "omp_critical", tcc_statement, 2)
 
 /* OpenACC - #pragma acc cache (variable1 ... variableN)
-   Operand 0: OACC_CACHE_CLAUSES: List of variables (transformed into
+   Operand 0: OMP_STANDALONE_CLAUSES: List of variables (transformed into
 	OMP_CLAUSE__CACHE_ clauses).  */
 DEFTREECODE (OACC_CACHE, "oacc_cache", tcc_statement, 1)
 
 /* OpenACC - #pragma acc declare [clause1 ... clauseN]
-   Operand 0: OACC_DECLARE_CLAUSES: List of clauses.  */
+   Operand 0: OMP_STANDALONE_CLAUSES: List of clauses.  */
 DEFTREECODE (OACC_DECLARE, "oacc_declare", tcc_statement, 1)
 
 /* OpenACC - #pragma acc enter data [clause1 ... clauseN]
-   Operand 0: OACC_ENTER_DATA_CLAUSES: List of clauses.  */
+   Operand 0: OMP_STANDALONE_CLAUSES: List of clauses.  */
 DEFTREECODE (OACC_ENTER_DATA, "oacc_enter_data", tcc_statement, 1)
 
 /* OpenACC - #pragma acc exit data [clause1 ... clauseN]
-   Operand 0: OACC_EXIT_DATA_CLAUSES: List of clauses.  */
+   Operand 0: OMP_STANDALONE_CLAUSES: List of clauses.  */
 DEFTREECODE (OACC_EXIT_DATA, "oacc_exit_data", tcc_statement, 1)
 
 /* OpenACC - #pragma acc update [clause1 ... clauseN]
-   Operand 0: OACC_UPDATE_CLAUSES: List of clauses.  */
+   Operand 0: OMP_STANDALONE_CLAUSES: List of clauses.  */
 DEFTREECODE (OACC_UPDATE, "oacc_update", tcc_statement, 1)
 
 /* OpenMP - #pragma omp target update [clause1 ... clauseN]
-   Operand 0: OMP_TARGET_UPDATE_CLAUSES: List of clauses.  */
+   Operand 0: OMP_STANDALONE_CLAUSES: List of clauses.  */
 DEFTREECODE (OMP_TARGET_UPDATE, "omp_target_update", tcc_statement, 1)
 
 /* OMP_ATOMIC through OMP_ATOMIC_CAPTURE_NEW must be consecutive,
diff --git gcc/tree.h gcc/tree.h
index 2ec9708..3c4633b 100644
--- gcc/tree.h
+++ gcc/tree.h
@@ -1222,21 +1222,6 @@ extern void protected_set_expr_location (tree, location_t);
 #define OACC_HOST_DATA_CLAUSES(NODE) \
   TREE_OPERAND (OACC_HOST_DATA_CHECK (NODE), 1)
 
-#define OACC_CACHE_CLAUSES(NODE) \
-  TREE_OPERAND (OACC_CACHE_CHECK (NODE), 0)
-
-#define OACC_DECLARE_CLAUSES(NODE) \
-  TREE_OPERAND (OACC_DECLARE_CHECK (NODE), 0)
-
-#define OACC_ENTER_DATA_CLAUSES(NODE) \
-  TREE_OPERAND (OACC_ENTER_DATA_CHECK (NODE), 0)
-
-#define OACC_EXIT_DATA_CLAUSES(NODE) \
-  TREE_OPERAND (OACC_EXIT_DATA_CHECK (NODE), 0)
-
-#define OACC_UPDATE_CLAUSES(NODE) \
-  TREE_OPERAND (OACC_UPDATE_CHECK (NODE), 0)
-
 #define OMP_PARALLEL_BODY(NODE)    TREE_OPERAND (OMP_PARALLEL_CHECK (NODE), 0)
 #define OMP_PARALLEL_CLAUSES(NODE) TREE_OPERAND (OMP_PARALLEL_CHECK (NODE), 1)
 
@@ -1283,8 +1268,8 @@ extern void protected_set_expr_location (tree, location_t);
 #define OMP_TARGET_BODY(NODE)	   TREE_OPERAND (OMP_TARGET_CHECK (NODE), 0)
 #define OMP_TARGET_CLAUSES(NODE)   TREE_OPERAND (OMP_TARGET_CHECK (NODE), 1)
 
-#define OMP_TARGET_UPDATE_CLAUSES(NODE)\
-  TREE_OPERAND (OMP_TARGET_UPDATE_CHECK (NODE), 0)
+#define OMP_STANDALONE_CLAUSES(NODE) \
+  TREE_OPERAND (TREE_RANGE_CHECK (NODE, OACC_CACHE, OMP_TARGET_UPDATE), 0)
 
 #define OMP_CLAUSE_SIZE(NODE)						\
   OMP_CLAUSE_OPERAND (OMP_CLAUSE_RANGE_CHECK (OMP_CLAUSE_CHECK (NODE),	\


Grüße,
 Thomas

[-- Attachment #2: Type: application/pgp-signature, Size: 472 bytes --]

Re: OMP_CLAUSES with clauses in operand 0

[Jakub Jelinek]

On Wed, Apr 29, 2015 at 04:31:32PM +0200, Thomas Schwinge wrote:
> > So yes, I really prefer OMP_STANDALONE_CLAUSES over OMP_CLAUSES for
> > everything.
> 
> Like this (for trunk)?
> 
> commit 300e28fce192cb56d73cb61f787872643030f0bf
> Author: Thomas Schwinge <thomas@codesourcery.com>
> Date:   Wed Apr 29 16:18:49 2015 +0200
> 
>     Add OMP_STANDALONE_CLAUSES.
>     
>     	gcc/
>     	* tree.h (OACC_CACHE_CLAUSES, OACC_DECLARE_CLAUSES)
>     	(OACC_ENTER_DATA_CLAUSES, OACC_EXIT_DATA_CLAUSES)
>     	(OACC_UPDATE_CLAUSES, OMP_TARGET_UPDATE_CLAUSES): Merge into...
>     	(OMP_STANDALONE_CLAUSES): ... this new macro.  Adjust all users.

I would keep the specific *_CLAUSES macros, just add
OMP_STANDALONE_CLAUSES and change the uses only if you are dealing with
multiple different codes.  That will match OMP_CLAUSES vs. OMP_FOR_CLAUSES,
OMP_PARALLEL_CLAUSES etc.

> --- gcc/c/c-parser.c
> +++ gcc/c/c-parser.c
> @@ -11987,7 +11987,7 @@ c_parser_oacc_cache (location_t loc, c_parser *parser)
>  
>    stmt = make_node (OACC_CACHE);
>    TREE_TYPE (stmt) = void_type_node;
> -  OACC_CACHE_CLAUSES (stmt) = clauses;
> +  OMP_STANDALONE_CLAUSES (stmt) = clauses;
>    SET_EXPR_LOCATION (stmt, loc);
>    add_stmt (stmt);
>  

So, drop hunks like this.

> @@ -12155,10 +12155,7 @@ c_parser_oacc_enter_exit_data (c_parser *parser, bool enter)
>  
>    stmt = enter ? make_node (OACC_ENTER_DATA) : make_node (OACC_EXIT_DATA);
>    TREE_TYPE (stmt) = void_type_node;
> -  if (enter)
> -    OACC_ENTER_DATA_CLAUSES (stmt) = clauses;
> -  else
> -    OACC_EXIT_DATA_CLAUSES (stmt) = clauses;
> +  OMP_STANDALONE_CLAUSES (stmt) = clauses;
>    SET_EXPR_LOCATION (stmt, loc);
>    add_stmt (stmt);
>  }

And just keep ones like this.

	Jakub

Re: OMP_CLAUSES with clauses in operand 0

[Thomas Schwinge]

[-- Attachment #1: Type: text/plain, Size: 6567 bytes --]

Hi Jakub!

On Wed, 29 Apr 2015 16:36:24 +0200, Jakub Jelinek <jakub@redhat.com> wrote:
> On Wed, Apr 29, 2015 at 04:31:32PM +0200, Thomas Schwinge wrote:
> > > So yes, I really prefer OMP_STANDALONE_CLAUSES over OMP_CLAUSES for
> > > everything.
> > 
> > Like this (for trunk)?
> > 
> > commit 300e28fce192cb56d73cb61f787872643030f0bf
> > Author: Thomas Schwinge <thomas@codesourcery.com>
> > Date:   Wed Apr 29 16:18:49 2015 +0200
> > 
> >     Add OMP_STANDALONE_CLAUSES.
> >     
> >     	gcc/
> >     	* tree.h (OACC_CACHE_CLAUSES, OACC_DECLARE_CLAUSES)
> >     	(OACC_ENTER_DATA_CLAUSES, OACC_EXIT_DATA_CLAUSES)
> >     	(OACC_UPDATE_CLAUSES, OMP_TARGET_UPDATE_CLAUSES): Merge into...
> >     	(OMP_STANDALONE_CLAUSES): ... this new macro.  Adjust all users.
> 
> I would keep the specific *_CLAUSES macros, just add
> OMP_STANDALONE_CLAUSES and change the uses only if you are dealing with
> multiple different codes.  That will match OMP_CLAUSES vs. OMP_FOR_CLAUSES,
> OMP_PARALLEL_CLAUSES etc.

My (non-explicit) rationale has been:

> > --- gcc/c/c-parser.c
> > +++ gcc/c/c-parser.c
> > @@ -11987,7 +11987,7 @@ c_parser_oacc_cache (location_t loc, c_parser *parser)
> >  
> >    stmt = make_node (OACC_CACHE);

We have just created a OACC_CACHE node here...

> >    TREE_TYPE (stmt) = void_type_node;
> > -  OACC_CACHE_CLAUSES (stmt) = clauses;
> > +  OMP_STANDALONE_CLAUSES (stmt) = clauses;

..., so there is no point in checking here that we're indeed dealing
specifically with an OACC_CACHE node.

> So, drop hunks like this.
> 
> > @@ -12155,10 +12155,7 @@ c_parser_oacc_enter_exit_data (c_parser *parser, bool enter)
> >  
> >    stmt = enter ? make_node (OACC_ENTER_DATA) : make_node (OACC_EXIT_DATA);
> >    TREE_TYPE (stmt) = void_type_node;
> > -  if (enter)
> > -    OACC_ENTER_DATA_CLAUSES (stmt) = clauses;
> > -  else
> > -    OACC_EXIT_DATA_CLAUSES (stmt) = clauses;
> > +  OMP_STANDALONE_CLAUSES (stmt) = clauses;
> >    SET_EXPR_LOCATION (stmt, loc);
> >    add_stmt (stmt);
> >  }
> 
> And just keep ones like this.

Done.  (I also reverted the gcc/cp/pt.c:tsubst_expr change which
motivated this patch; will include that with the patch adding support for
C++ templates usage with OpenACC directives.)  OK for trunk?

commit 82e588b6d62f9e7254e76a3dfcc46efceb2075a5
Author: Thomas Schwinge <thomas@codesourcery.com>
Date:   Wed Apr 29 17:08:17 2015 +0200

    Add OMP_STANDALONE_CLAUSES.
    
    	gcc/
    	* tree.h (OMP_STANDALONE_CLAUSES): New macro.
    	* gimplify.c (gimplify_omp_workshare): Use it.
    	gcc/c/
    	* c-parser.c (c_parser_oacc_enter_exit_data): Use
    	OMP_STANDALONE_CLAUSES.
    	gcc/cp/
    	* parser.c (cp_parser_oacc_enter_exit_data): Use
    	OMP_STANDALONE_CLAUSES.
---
 gcc/c/c-parser.c |    5 +----
 gcc/cp/parser.c  |    5 +----
 gcc/gimplify.c   |   13 +++++--------
 gcc/tree.h       |    6 ++++++
 4 files changed, 13 insertions(+), 16 deletions(-)

diff --git gcc/c/c-parser.c gcc/c/c-parser.c
index f5e2ac2c..015de7f 100644
--- gcc/c/c-parser.c
+++ gcc/c/c-parser.c
@@ -12155,10 +12155,7 @@ c_parser_oacc_enter_exit_data (c_parser *parser, bool enter)
 
   stmt = enter ? make_node (OACC_ENTER_DATA) : make_node (OACC_EXIT_DATA);
   TREE_TYPE (stmt) = void_type_node;
-  if (enter)
-    OACC_ENTER_DATA_CLAUSES (stmt) = clauses;
-  else
-    OACC_EXIT_DATA_CLAUSES (stmt) = clauses;
+  OMP_STANDALONE_CLAUSES (stmt) = clauses;
   SET_EXPR_LOCATION (stmt, loc);
   add_stmt (stmt);
 }
diff --git gcc/cp/parser.c gcc/cp/parser.c
index 4ea2ca2..cfb512b 100644
--- gcc/cp/parser.c
+++ gcc/cp/parser.c
@@ -31606,10 +31606,7 @@ cp_parser_oacc_enter_exit_data (cp_parser *parser, cp_token *pragma_tok,
 
   stmt = enter ? make_node (OACC_ENTER_DATA) : make_node (OACC_EXIT_DATA);
   TREE_TYPE (stmt) = void_type_node;
-  if (enter)
-    OACC_ENTER_DATA_CLAUSES (stmt) = clauses;
-  else
-    OACC_EXIT_DATA_CLAUSES (stmt) = clauses;
+  OMP_STANDALONE_CLAUSES (stmt) = clauses;
   SET_EXPR_LOCATION (stmt, pragma_tok->location);
   add_stmt (stmt);
   return stmt;
diff --git gcc/gimplify.c gcc/gimplify.c
index c68bd47..bda62ce 100644
--- gcc/gimplify.c
+++ gcc/gimplify.c
@@ -7427,34 +7427,31 @@ gimplify_omp_workshare (tree *expr_p, gimple_seq *pre_p)
 static void
 gimplify_omp_target_update (tree *expr_p, gimple_seq *pre_p)
 {
-  tree expr = *expr_p, clauses;
+  tree expr = *expr_p;
   int kind;
   gomp_target *stmt;
 
   switch (TREE_CODE (expr))
     {
     case OACC_ENTER_DATA:
-      clauses = OACC_ENTER_DATA_CLAUSES (expr);
       kind = GF_OMP_TARGET_KIND_OACC_ENTER_EXIT_DATA;
       break;
     case OACC_EXIT_DATA:
-      clauses = OACC_EXIT_DATA_CLAUSES (expr);
       kind = GF_OMP_TARGET_KIND_OACC_ENTER_EXIT_DATA;
       break;
     case OACC_UPDATE:
-      clauses = OACC_UPDATE_CLAUSES (expr);
       kind = GF_OMP_TARGET_KIND_OACC_UPDATE;
       break;
     case OMP_TARGET_UPDATE:
-      clauses = OMP_TARGET_UPDATE_CLAUSES (expr);
       kind = GF_OMP_TARGET_KIND_UPDATE;
       break;
     default:
       gcc_unreachable ();
     }
-  gimplify_scan_omp_clauses (&clauses, pre_p, ORT_WORKSHARE);
-  gimplify_adjust_omp_clauses (pre_p, &clauses);
-  stmt = gimple_build_omp_target (NULL, kind, clauses);
+  gimplify_scan_omp_clauses (&OMP_STANDALONE_CLAUSES (expr), pre_p,
+			     ORT_WORKSHARE);
+  gimplify_adjust_omp_clauses (pre_p, &OMP_STANDALONE_CLAUSES (expr));
+  stmt = gimple_build_omp_target (NULL, kind, OMP_STANDALONE_CLAUSES (expr));
 
   gimplify_seq_add_stmt (pre_p, stmt);
   *expr_p = NULL_TREE;
diff --git gcc/tree.h gcc/tree.h
index 2ec9708..e17bd9b 100644
--- gcc/tree.h
+++ gcc/tree.h
@@ -1197,11 +1197,17 @@ extern void protected_set_expr_location (tree, location_t);
 
 /* OpenMP and OpenACC directive and clause accessors.  */
 
+/* Generic accessors for OMP nodes that keep the body as operand 0, and clauses
+   as operand 1.  */
 #define OMP_BODY(NODE) \
   TREE_OPERAND (TREE_RANGE_CHECK (NODE, OACC_PARALLEL, OMP_CRITICAL), 0)
 #define OMP_CLAUSES(NODE) \
   TREE_OPERAND (TREE_RANGE_CHECK (NODE, OACC_PARALLEL, OMP_SINGLE), 1)
 
+/* Generic accessors for OMP nodes that keep clauses as operand 0.  */
+#define OMP_STANDALONE_CLAUSES(NODE) \
+  TREE_OPERAND (TREE_RANGE_CHECK (NODE, OACC_CACHE, OMP_TARGET_UPDATE), 0)
+
 #define OACC_PARALLEL_BODY(NODE) \
   TREE_OPERAND (OACC_PARALLEL_CHECK (NODE), 0)
 #define OACC_PARALLEL_CLAUSES(NODE) \


Grüße,
 Thomas

[-- Attachment #2: Type: application/pgp-signature, Size: 472 bytes --]

Re: OMP_CLAUSES with clauses in operand 0

[Jakub Jelinek]

On Wed, Apr 29, 2015 at 05:25:39PM +0200, Thomas Schwinge wrote:
> Done.  (I also reverted the gcc/cp/pt.c:tsubst_expr change which
> motivated this patch; will include that with the patch adding support for
> C++ templates usage with OpenACC directives.)  OK for trunk?

Ok, thanks.

	Jakub

Re: OMP_CLAUSES with clauses in operand 0

[Thomas Schwinge]

[-- Attachment #1: Type: text/plain, Size: 6052 bytes --]

Hi!

On Wed, 29 Apr 2015 17:28:54 +0200, Jakub Jelinek <jakub@redhat.com> wrote:
> On Wed, Apr 29, 2015 at 05:25:39PM +0200, Thomas Schwinge wrote:
> > Done.  (I also reverted the gcc/cp/pt.c:tsubst_expr change which
> > motivated this patch; will include that with the patch adding support for
> > C++ templates usage with OpenACC directives.)  OK for trunk?
> 
> Ok, thanks.

Committed in r222580:

commit 95cfd3918aef02196f24de30a1e7cbd34e45e827
Author: tschwinge <tschwinge@138bc75d-0d04-0410-961f-82ee72b054a4>
Date:   Wed Apr 29 15:44:41 2015 +0000

    Add OMP_STANDALONE_CLAUSES.
    
    	gcc/
    	* tree.h (OMP_STANDALONE_CLAUSES): New macro.
    	* gimplify.c (gimplify_omp_workshare): Use it.
    	gcc/c/
    	* c-parser.c (c_parser_oacc_enter_exit_data): Use
    	OMP_STANDALONE_CLAUSES.
    	gcc/cp/
    	* parser.c (cp_parser_oacc_enter_exit_data): Use
    	OMP_STANDALONE_CLAUSES.
    
    git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@222580 138bc75d-0d04-0410-961f-82ee72b054a4
---
 gcc/ChangeLog    |    5 +++++
 gcc/c/ChangeLog  |    5 +++++
 gcc/c/c-parser.c |    5 +----
 gcc/cp/ChangeLog |    5 +++++
 gcc/cp/parser.c  |    5 +----
 gcc/gimplify.c   |   13 +++++--------
 gcc/tree.h       |    6 ++++++
 7 files changed, 28 insertions(+), 16 deletions(-)

diff --git gcc/ChangeLog gcc/ChangeLog
index 4fb5490..11cb62a 100644
--- gcc/ChangeLog
+++ gcc/ChangeLog
@@ -1,3 +1,8 @@
+2015-04-29  Thomas Schwinge  <thomas@codesourcery.com>
+
+	* tree.h (OMP_STANDALONE_CLAUSES): New macro.
+	* gimplify.c (gimplify_omp_workshare): Use it.
+
 2015-04-29  Richard Sandiford  <richard.sandiford@arm.com>
 
 	* Makefile.in (build/genrecog.o): Depend on inchash.h.
diff --git gcc/c/ChangeLog gcc/c/ChangeLog
index 9c769ca..6d8dbb1 100644
--- gcc/c/ChangeLog
+++ gcc/c/ChangeLog
@@ -1,3 +1,8 @@
+2015-04-29  Thomas Schwinge  <thomas@codesourcery.com>
+
+	* c-parser.c (c_parser_oacc_enter_exit_data): Use
+	OMP_STANDALONE_CLAUSES.
+
 2015-04-28  Marek Polacek  <polacek@redhat.com>
 
 	* c-parser.c (c_parser_binary_expression): Remove duplicate line.
diff --git gcc/c/c-parser.c gcc/c/c-parser.c
index cc8a4e3..bf0e4c57 100644
--- gcc/c/c-parser.c
+++ gcc/c/c-parser.c
@@ -12154,10 +12154,7 @@ c_parser_oacc_enter_exit_data (c_parser *parser, bool enter)
 
   stmt = enter ? make_node (OACC_ENTER_DATA) : make_node (OACC_EXIT_DATA);
   TREE_TYPE (stmt) = void_type_node;
-  if (enter)
-    OACC_ENTER_DATA_CLAUSES (stmt) = clauses;
-  else
-    OACC_EXIT_DATA_CLAUSES (stmt) = clauses;
+  OMP_STANDALONE_CLAUSES (stmt) = clauses;
   SET_EXPR_LOCATION (stmt, loc);
   add_stmt (stmt);
 }
diff --git gcc/cp/ChangeLog gcc/cp/ChangeLog
index 3ee050c..9442faa 100644
--- gcc/cp/ChangeLog
+++ gcc/cp/ChangeLog
@@ -1,3 +1,8 @@
+2015-04-29  Thomas Schwinge  <thomas@codesourcery.com>
+
+	* parser.c (cp_parser_oacc_enter_exit_data): Use
+	OMP_STANDALONE_CLAUSES.
+
 2015-04-29  Paolo Carlini  <paolo.carlini@oracle.com>
 
 	PR c++/64667
diff --git gcc/cp/parser.c gcc/cp/parser.c
index 4ea2ca2..cfb512b 100644
--- gcc/cp/parser.c
+++ gcc/cp/parser.c
@@ -31606,10 +31606,7 @@ cp_parser_oacc_enter_exit_data (cp_parser *parser, cp_token *pragma_tok,
 
   stmt = enter ? make_node (OACC_ENTER_DATA) : make_node (OACC_EXIT_DATA);
   TREE_TYPE (stmt) = void_type_node;
-  if (enter)
-    OACC_ENTER_DATA_CLAUSES (stmt) = clauses;
-  else
-    OACC_EXIT_DATA_CLAUSES (stmt) = clauses;
+  OMP_STANDALONE_CLAUSES (stmt) = clauses;
   SET_EXPR_LOCATION (stmt, pragma_tok->location);
   add_stmt (stmt);
   return stmt;
diff --git gcc/gimplify.c gcc/gimplify.c
index 1d5341e..9ce3dd9 100644
--- gcc/gimplify.c
+++ gcc/gimplify.c
@@ -7411,34 +7411,31 @@ gimplify_omp_workshare (tree *expr_p, gimple_seq *pre_p)
 static void
 gimplify_omp_target_update (tree *expr_p, gimple_seq *pre_p)
 {
-  tree expr = *expr_p, clauses;
+  tree expr = *expr_p;
   int kind;
   gomp_target *stmt;
 
   switch (TREE_CODE (expr))
     {
     case OACC_ENTER_DATA:
-      clauses = OACC_ENTER_DATA_CLAUSES (expr);
       kind = GF_OMP_TARGET_KIND_OACC_ENTER_EXIT_DATA;
       break;
     case OACC_EXIT_DATA:
-      clauses = OACC_EXIT_DATA_CLAUSES (expr);
       kind = GF_OMP_TARGET_KIND_OACC_ENTER_EXIT_DATA;
       break;
     case OACC_UPDATE:
-      clauses = OACC_UPDATE_CLAUSES (expr);
       kind = GF_OMP_TARGET_KIND_OACC_UPDATE;
       break;
     case OMP_TARGET_UPDATE:
-      clauses = OMP_TARGET_UPDATE_CLAUSES (expr);
       kind = GF_OMP_TARGET_KIND_UPDATE;
       break;
     default:
       gcc_unreachable ();
     }
-  gimplify_scan_omp_clauses (&clauses, pre_p, ORT_WORKSHARE);
-  gimplify_adjust_omp_clauses (pre_p, &clauses);
-  stmt = gimple_build_omp_target (NULL, kind, clauses);
+  gimplify_scan_omp_clauses (&OMP_STANDALONE_CLAUSES (expr), pre_p,
+			     ORT_WORKSHARE);
+  gimplify_adjust_omp_clauses (pre_p, &OMP_STANDALONE_CLAUSES (expr));
+  stmt = gimple_build_omp_target (NULL, kind, OMP_STANDALONE_CLAUSES (expr));
 
   gimplify_seq_add_stmt (pre_p, stmt);
   *expr_p = NULL_TREE;
diff --git gcc/tree.h gcc/tree.h
index 2ec9708..e17bd9b 100644
--- gcc/tree.h
+++ gcc/tree.h
@@ -1197,11 +1197,17 @@ extern void protected_set_expr_location (tree, location_t);
 
 /* OpenMP and OpenACC directive and clause accessors.  */
 
+/* Generic accessors for OMP nodes that keep the body as operand 0, and clauses
+   as operand 1.  */
 #define OMP_BODY(NODE) \
   TREE_OPERAND (TREE_RANGE_CHECK (NODE, OACC_PARALLEL, OMP_CRITICAL), 0)
 #define OMP_CLAUSES(NODE) \
   TREE_OPERAND (TREE_RANGE_CHECK (NODE, OACC_PARALLEL, OMP_SINGLE), 1)
 
+/* Generic accessors for OMP nodes that keep clauses as operand 0.  */
+#define OMP_STANDALONE_CLAUSES(NODE) \
+  TREE_OPERAND (TREE_RANGE_CHECK (NODE, OACC_CACHE, OMP_TARGET_UPDATE), 0)
+
 #define OACC_PARALLEL_BODY(NODE) \
   TREE_OPERAND (OACC_PARALLEL_CHECK (NODE), 0)
 #define OACC_PARALLEL_CLAUSES(NODE) \


Grüße,
 Thomas

[-- Attachment #2: Type: application/pgp-signature, Size: 472 bytes --]