[gomp4] OpenACC cache directive maintenance (was: [PATCH 4/6] [GOMP4] OpenACC 1.0+ support in fortran front-end)

[Thomas Schwinge <thomas@codesourcery.com>]

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

Hi!

On Fri, 24 Jan 2014 20:33:35 +0100, I wrote:
> On Thu, 23 Jan 2014 22:04:45 +0400, Ilmir Usmanov <i.usmanov@samsung.com> wrote:
> > Subject: [PATCH 4/6] OpenACC GENERIC nodes

> > --- a/gcc/tree-core.h
> > +++ b/gcc/tree-core.h
> > @@ -216,12 +216,18 @@ enum omp_clause_code {

> > +  /* Internal structure to hold OpenACC cache directive's variable-list.
> > +     #pragma acc cache (variable-_ist).  */
> > +  OACC_NO_CLAUSE_CACHE,
> 
> Hmm, yeah, while *_NO_CLAUSE_* perhaps isn't the most beautiful approach,
> I think it's fine at least for now.

In r217146, I applied the following to gomp-4_0-branch:

commit e8e44b733808997d06c0cdf9bf5756ce03530f42
Author: tschwinge <tschwinge@138bc75d-0d04-0410-961f-82ee72b054a4>
Date:   Wed Nov 5 16:35:30 2014 +0000

    OpenACC cache directive maintenance.
    
    	gcc/c/
    	* c-parser.c (c_parser_oacc_cache): Generate OACC_CACHE.
    	* c-typeck.c (c_finish_omp_clauses): Handle OMP_CLAUSE__CACHE_.
    	gcc/cp/
    	* parser.c (cp_parser_oacc_cache): Generate OACC_CACHE.
    	* semantics.c (finish_omp_clauses): Handle OMP_CLAUSE__CACHE_.
    	gcc/
    	* gimplify.c (gimplify_oacc_cache): New function.
    	(gimplify_expr): Use it for OACC_CACHE.
    	(gimplify_scan_omp_clauses, gimplify_adjust_omp_clauses): Handle
    	OMP_CLAUSE__CACHE_.
    
    	gcc/c/
    	* c-parser.c (c_parser_omp_variable_list) <OMP_CLAUSE__CACHE_>:
    	Remove explicit mark_exp_read invocations.
    	gcc/cp/
    	* parser.c (cp_parser_omp_var_list_no_open) <OMP_CLAUSE__CACHE_>:
    	Remove explicit mark_exp_read invocations.
    
    	gcc/
    	* tree-core.h (enum omp_clause_code): Move OMP_NO_CLAUSE_CACHE
    	next to, and handle it like a data clause.  Rename it to
    	OMP_CLAUSE__CACHE_.  Update all users.
    
    git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/branches/gomp-4_0-branch@217146 138bc75d-0d04-0410-961f-82ee72b054a4
---
 gcc/ChangeLog.gomp         |  9 +++++++++
 gcc/c/ChangeLog.gomp       |  8 ++++++++
 gcc/c/c-parser.c           | 23 +++++++++++++++--------
 gcc/c/c-typeck.c           |  1 +
 gcc/cp/ChangeLog.gomp      |  6 ++++++
 gcc/cp/parser.c            | 24 +++++++++++++++---------
 gcc/cp/semantics.c         |  1 +
 gcc/fortran/trans-openmp.c |  2 +-
 gcc/gimplify.c             | 25 ++++++++++++++++++++++---
 gcc/omp-low.c              |  4 ++--
 gcc/tree-core.h            |  8 ++++----
 gcc/tree-pretty-print.c    | 11 +++++++----
 gcc/tree.c                 |  6 +++---
 gcc/tree.def               |  5 +++--
 gcc/tree.h                 |  2 +-
 15 files changed, 98 insertions(+), 37 deletions(-)

diff --git gcc/ChangeLog.gomp gcc/ChangeLog.gomp
index fc624c8..2c2b349 100644
--- gcc/ChangeLog.gomp
+++ gcc/ChangeLog.gomp
@@ -1,5 +1,14 @@
 2014-11-05  Thomas Schwinge  <thomas@codesourcery.com>
 
+	* gimplify.c (gimplify_oacc_cache): New function.
+	(gimplify_expr): Use it for OACC_CACHE.
+	(gimplify_scan_omp_clauses, gimplify_adjust_omp_clauses): Handle
+	OMP_CLAUSE__CACHE_.
+
+	* tree-core.h (enum omp_clause_code): Move OMP_NO_CLAUSE_CACHE
+	next to, and handle it like a data clause.  Rename it to
+	OMP_CLAUSE__CACHE_.  Update all users.
+
 	* invoke.texi: Update for OpenACC.
 	* sourcebuild.texi: Likewise.
 
diff --git gcc/c/ChangeLog.gomp gcc/c/ChangeLog.gomp
index 7acd7b3..70278b9 100644
--- gcc/c/ChangeLog.gomp
+++ gcc/c/ChangeLog.gomp
@@ -1,3 +1,11 @@
+2014-11-05  Thomas Schwinge  <thomas@codesourcery.com>
+
+	* c-parser.c (c_parser_oacc_cache): Generate OACC_CACHE.
+	* c-typeck.c (c_finish_omp_clauses): Handle OMP_CLAUSE__CACHE_.
+
+	* c-parser.c (c_parser_omp_variable_list) <OMP_CLAUSE__CACHE_>:
+	Remove explicit mark_exp_read invocations.
+
 2014-11-05  James Norris  <jnorris@codesourcery.com>
 
 	* c-parser.c (c_parser_omp_variable_list): Handle
diff --git gcc/c/c-parser.c gcc/c/c-parser.c
index 410b19f..40d4314 100644
--- gcc/c/c-parser.c
+++ gcc/c/c-parser.c
@@ -10053,7 +10053,7 @@ c_parser_omp_variable_list (c_parser *parser,
 	{
 	  switch (kind)
 	    {
-	    case OMP_NO_CLAUSE_CACHE:
+	    case OMP_CLAUSE__CACHE_:
 	      if (c_parser_peek_token (parser)->type != CPP_OPEN_SQUARE)
 		{
 		  c_parser_error (parser, "expected %<[%>");
@@ -10100,11 +10100,8 @@ c_parser_omp_variable_list (c_parser *parser,
 		      break;
 		    }
 
-		  if (kind == OMP_NO_CLAUSE_CACHE)
+		  if (kind == OMP_CLAUSE__CACHE_)
 		    {
-		      mark_exp_read (low_bound);
-		      mark_exp_read (length);
-
 		      if (TREE_CODE (low_bound) != INTEGER_CST
 			  && !TREE_READONLY (low_bound))
 			{
@@ -11901,12 +11898,22 @@ c_parser_omp_structured_block (c_parser *parser)
 */
 
 static tree
-c_parser_oacc_cache (location_t loc __attribute__((unused)), c_parser *parser)
+c_parser_oacc_cache (location_t loc, c_parser *parser)
 {
-  c_parser_omp_var_list_parens (parser, OMP_NO_CLAUSE_CACHE, NULL);
+  tree stmt, clauses;
+
+  clauses = c_parser_omp_var_list_parens (parser, OMP_CLAUSE__CACHE_, NULL);
+  clauses = c_finish_omp_clauses (clauses);
+
   c_parser_skip_to_pragma_eol (parser);
 
-  return NULL_TREE;
+  stmt = make_node (OACC_CACHE);
+  TREE_TYPE (stmt) = void_type_node;
+  OACC_CACHE_CLAUSES (stmt) = clauses;
+  SET_EXPR_LOCATION (stmt, loc);
+  add_stmt (stmt);
+
+  return stmt;
 }
 
 /* OpenACC 2.0:
diff --git gcc/c/c-typeck.c gcc/c/c-typeck.c
index 76503e4..e315690d 100644
--- gcc/c/c-typeck.c
+++ gcc/c/c-typeck.c
@@ -12204,6 +12204,7 @@ c_finish_omp_clauses (tree clauses)
 	case OMP_CLAUSE_MAP:
 	case OMP_CLAUSE_TO:
 	case OMP_CLAUSE_FROM:
+	case OMP_CLAUSE__CACHE_:
 	  t = OMP_CLAUSE_DECL (c);
 	  if (TREE_CODE (t) == TREE_LIST)
 	    {
diff --git gcc/cp/ChangeLog.gomp gcc/cp/ChangeLog.gomp
index 024e6a5..46d4912 100644
--- gcc/cp/ChangeLog.gomp
+++ gcc/cp/ChangeLog.gomp
@@ -1,5 +1,11 @@
 2014-11-05  Thomas Schwinge  <thomas@codesourcery.com>
 
+	* parser.c (cp_parser_oacc_cache): Generate OACC_CACHE.
+	* semantics.c (finish_omp_clauses): Handle OMP_CLAUSE__CACHE_.
+
+	* parser.c (cp_parser_omp_var_list_no_open) <OMP_CLAUSE__CACHE_>:
+	Remove explicit mark_exp_read invocations.
+
 	* parser.c (cp_parser_omp_clause_name): Also look for "pcopy",
 	"pcopyin", "pcopyout", "pcreate".  Look for "wait" instead of
 	"WAIT".
diff --git gcc/cp/parser.c gcc/cp/parser.c
index 3ef2de7..ea4ad2f 100644
--- gcc/cp/parser.c
+++ gcc/cp/parser.c
@@ -27669,7 +27669,7 @@ cp_parser_omp_var_list_no_open (cp_parser *parser, enum omp_clause_code kind,
 	{
 	  switch (kind)
 	    {
-	    case OMP_NO_CLAUSE_CACHE:
+	    case OMP_CLAUSE__CACHE_:
 	      if (cp_lexer_peek_token (parser->lexer)->type != CPP_OPEN_SQUARE)
 		{
 		  error_at (token->location, "expected %<[%>");
@@ -27708,11 +27708,8 @@ cp_parser_omp_var_list_no_open (cp_parser *parser, enum omp_clause_code kind,
 					  RT_CLOSE_SQUARE))
 		    goto skip_comma;
 
-		  if (kind == OMP_NO_CLAUSE_CACHE)
+		  if (kind == OMP_CLAUSE__CACHE_)
 		    {
-		      mark_exp_read (low_bound);
-		      mark_exp_read (length);
-
 		      if (TREE_CODE (low_bound) != INTEGER_CST
 			  && !TREE_READONLY (low_bound))
 			{
@@ -31410,13 +31407,22 @@ cp_parser_omp_target (cp_parser *parser, cp_token *pragma_tok,
 */
 
 static tree
-cp_parser_oacc_cache (cp_parser *parser,
-				cp_token *pragma_tok __attribute__((unused)))
+cp_parser_oacc_cache (cp_parser *parser, cp_token *pragma_tok)
 {
-  cp_parser_omp_var_list (parser, OMP_NO_CLAUSE_CACHE, NULL_TREE);
+  tree stmt, clauses;
+
+  clauses = cp_parser_omp_var_list (parser, OMP_CLAUSE__CACHE_, NULL_TREE);
+  clauses = finish_omp_clauses (clauses);
+
   cp_parser_require_pragma_eol (parser, cp_lexer_peek_token (parser->lexer));
 
-  return NULL_TREE;
+  stmt = make_node (OACC_CACHE);
+  TREE_TYPE (stmt) = void_type_node;
+  OACC_CACHE_CLAUSES (stmt) = clauses;
+  SET_EXPR_LOCATION (stmt, pragma_tok->location);
+  add_stmt (stmt);
+
+  return stmt;
 }
 
 /* OpenACC 2.0:
diff --git gcc/cp/semantics.c gcc/cp/semantics.c
index 2457a6f..6e35eef 100644
--- gcc/cp/semantics.c
+++ gcc/cp/semantics.c
@@ -5704,6 +5704,7 @@ finish_omp_clauses (tree clauses)
 	case OMP_CLAUSE_MAP:
 	case OMP_CLAUSE_TO:
 	case OMP_CLAUSE_FROM:
+	case OMP_CLAUSE__CACHE_:
 	  t = OMP_CLAUSE_DECL (c);
 	  if (TREE_CODE (t) == TREE_LIST)
 	    {
diff --git gcc/fortran/trans-openmp.c gcc/fortran/trans-openmp.c
index 97613ae..7dd4498 100644
--- gcc/fortran/trans-openmp.c
+++ gcc/fortran/trans-openmp.c
@@ -1807,7 +1807,7 @@ gfc_trans_omp_clauses (stmtblock_t *block, gfc_omp_clauses *clauses,
 	  clause_code = OMP_CLAUSE_DEVICE_RESIDENT;
 	  goto add_clause;
 	case OMP_LIST_CACHE:
-	  clause_code = OMP_NO_CLAUSE_CACHE;
+	  clause_code = OMP_CLAUSE__CACHE_;
 	  goto add_clause;
 
 	add_clause:
diff --git gcc/gimplify.c gcc/gimplify.c
index bfd7f66..d58876f 100644
--- gcc/gimplify.c
+++ gcc/gimplify.c
@@ -6114,6 +6114,7 @@ gimplify_scan_omp_clauses (tree *list_p, gimple_seq *pre_p,
 
 	case OMP_CLAUSE_TO:
 	case OMP_CLAUSE_FROM:
+	case OMP_CLAUSE__CACHE_:
 	  decl = OMP_CLAUSE_DECL (c);
 	  if (error_operand_p (decl))
 	    {
@@ -6294,7 +6295,6 @@ gimplify_scan_omp_clauses (tree *list_p, gimple_seq *pre_p,
 	case OMP_CLAUSE_GANG:
 	case OMP_CLAUSE_ASYNC:
 	case OMP_CLAUSE_WAIT:
-	case OMP_NO_CLAUSE_CACHE:
 	case OMP_CLAUSE_INDEPENDENT:
 	case OMP_CLAUSE_WORKER:
 	case OMP_CLAUSE_VECTOR:
@@ -6641,6 +6641,7 @@ gimplify_adjust_omp_clauses (gimple_seq *pre_p, tree *list_p)
 
 	case OMP_CLAUSE_TO:
 	case OMP_CLAUSE_FROM:
+	case OMP_CLAUSE__CACHE_:
 	  decl = OMP_CLAUSE_DECL (c);
 	  if (!DECL_P (decl))
 	    break;
@@ -6698,7 +6699,6 @@ gimplify_adjust_omp_clauses (gimple_seq *pre_p, tree *list_p)
 	case OMP_CLAUSE_GANG:
 	case OMP_CLAUSE_ASYNC:
 	case OMP_CLAUSE_WAIT:
-	case OMP_NO_CLAUSE_CACHE:
 	case OMP_CLAUSE_INDEPENDENT:
 	case OMP_CLAUSE_WORKER:
 	case OMP_CLAUSE_VECTOR:
@@ -6722,6 +6722,21 @@ gimplify_adjust_omp_clauses (gimple_seq *pre_p, tree *list_p)
   delete_omp_context (ctx);
 }
 
+/* Gimplify OACC_CACHE.  */
+
+static void
+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));
+
+  /* TODO: Do something sensible with this information.  */
+
+  *expr_p = NULL_TREE;
+}
+
 /* Gimplify the contents of an OMP_PARALLEL statement.  This involves
    gimplification of the body, as well as scanning the body for used
    variables.  We need to do this scan now, because variable-sized
@@ -8312,7 +8327,6 @@ gimplify_expr (tree *expr_p, gimple_seq *pre_p, gimple_seq *post_p,
 
 	case OACC_HOST_DATA:
 	case OACC_DECLARE:
-	case OACC_CACHE:
 	  sorry ("directive not yet implemented");
 	  ret = GS_ALL_DONE;
 	  break;
@@ -8352,6 +8366,11 @@ gimplify_expr (tree *expr_p, gimple_seq *pre_p, gimple_seq *post_p,
 	  ret = GS_ALL_DONE;
 	  break;
 
+	case OACC_CACHE:
+	  gimplify_oacc_cache (expr_p, pre_p);
+	  ret = GS_ALL_DONE;
+	  break;
+
 	case OACC_DATA:
 	case OMP_SECTIONS:
 	case OMP_SINGLE:
diff --git gcc/omp-low.c gcc/omp-low.c
index 49cf1ab..1c9d942 100644
--- gcc/omp-low.c
+++ gcc/omp-low.c
@@ -1982,7 +1982,7 @@ scan_sharing_clauses (tree clauses, omp_context *ctx)
 	case OMP_CLAUSE_DEVICE_RESIDENT:
 	case OMP_CLAUSE_USE_DEVICE:
 	case OMP_CLAUSE_GANG:
-	case OMP_NO_CLAUSE_CACHE:
+	case OMP_CLAUSE__CACHE_:
 	case OMP_CLAUSE_INDEPENDENT:
 	case OMP_CLAUSE_WORKER:
 	case OMP_CLAUSE_VECTOR:
@@ -2130,7 +2130,7 @@ scan_sharing_clauses (tree clauses, omp_context *ctx)
 	case OMP_CLAUSE_DEVICE_RESIDENT:
 	case OMP_CLAUSE_USE_DEVICE:
 	case OMP_CLAUSE_GANG:
-	case OMP_NO_CLAUSE_CACHE:
+	case OMP_CLAUSE__CACHE_:
 	case OMP_CLAUSE_INDEPENDENT:
 	case OMP_CLAUSE_WORKER:
 	case OMP_CLAUSE_VECTOR:
diff --git gcc/tree-core.h gcc/tree-core.h
index abdc2c9..42ad6a0 100644
--- gcc/tree-core.h
+++ gcc/tree-core.h
@@ -266,6 +266,10 @@ enum omp_clause_code {
      OpenMP clause: map ({alloc:,to:,from:,tofrom:,}variable-list).  */
   OMP_CLAUSE_MAP,
 
+  /* Internal structure to hold OpenACC cache directive's variable-list.
+     #pragma acc cache (variable-list).  */
+  OMP_CLAUSE__CACHE_,
+
   /* OpenACC clause: host (variable_list).  */
   OMP_CLAUSE_HOST,
 
@@ -292,10 +296,6 @@ enum omp_clause_code {
   /* OpenACC clause/directive: wait [(integer-expression-list)].  */
   OMP_CLAUSE_WAIT,
 
-  /* Internal structure to hold OpenACC cache directive's variable-list.
-     #pragma acc cache (variable-list).  */
-  OMP_NO_CLAUSE_CACHE,
-
   /* Internal clause: temporary for combined loops expansion.  */
   OMP_CLAUSE__LOOPTEMP_,
 
diff --git gcc/tree-pretty-print.c gcc/tree-pretty-print.c
index f311ed9..d678f36 100644
--- gcc/tree-pretty-print.c
+++ gcc/tree-pretty-print.c
@@ -347,9 +347,6 @@ dump_omp_clause (pretty_printer *buffer, tree clause, int spc, int flags)
     case OMP_CLAUSE_USE_DEVICE:
       name = "use_device";
       goto print_remap;
-    case OMP_NO_CLAUSE_CACHE:
-      name = "_cache_";
-      goto print_remap;
   print_remap:
       pp_string (buffer, name);
       pp_left_paren (buffer);
@@ -599,6 +596,12 @@ dump_omp_clause (pretty_printer *buffer, tree clause, int spc, int flags)
 			 spc, flags, false);
       goto print_clause_size;
 
+    case OMP_CLAUSE__CACHE_:
+      pp_string (buffer, "(");
+      dump_generic_node (buffer, OMP_CLAUSE_DECL (clause),
+			 spc, flags, false);
+      goto print_clause_size;
+
     case OMP_CLAUSE_NUM_TEAMS:
       pp_string (buffer, "num_teams(");
       dump_generic_node (buffer, OMP_CLAUSE_NUM_TEAMS_EXPR (clause),
@@ -2548,7 +2551,7 @@ dump_generic_node (pretty_printer *buffer, tree node, int spc, int flags,
 
     case OACC_CACHE:
       pp_string (buffer, "#pragma acc cache");
-      dump_omp_clauses (buffer, OACC_CACHE_CLAUSES(node), spc, flags);
+      dump_omp_clauses (buffer, OACC_CACHE_CLAUSES (node), spc, flags);
       break;
 
     case OMP_PARALLEL:
diff --git gcc/tree.c gcc/tree.c
index 0475622..f39c63f 100644
--- gcc/tree.c
+++ gcc/tree.c
@@ -270,6 +270,7 @@ unsigned const char omp_clause_num_ops[] =
   2, /* OMP_CLAUSE_FROM  */
   2, /* OMP_CLAUSE_TO  */
   2, /* OMP_CLAUSE_MAP  */
+  2, /* OMP_CLAUSE__CACHE_  */
   1, /* OMP_CLAUSE_HOST  */
   1, /* OMP_CLAUSE_OACC_DEVICE  */
   1, /* OMP_CLAUSE_DEVICE_RESIDENT  */
@@ -277,7 +278,6 @@ unsigned const char omp_clause_num_ops[] =
   1, /* OMP_CLAUSE_GANG  */
   1, /* OMP_CLAUSE_ASYNC  */
   1, /* OMP_CLAUSE_WAIT  */
-  1, /* OMP_NO_CLAUSE_CACHE  */
   1, /* OMP_CLAUSE__LOOPTEMP_  */
   1, /* OMP_CLAUSE_IF  */
   1, /* OMP_CLAUSE_NUM_THREADS  */
@@ -329,6 +329,7 @@ const char * const omp_clause_code_name[] =
   "from",
   "to",
   "map",
+  "_cache_",
   "host",
   "device",
   "device_resident",
@@ -336,7 +337,6 @@ const char * const omp_clause_code_name[] =
   "gang",
   "async",
   "wait",
-  "_cache_",
   "_looptemp_",
   "if",
   "num_threads",
@@ -11127,7 +11127,6 @@ walk_tree_1 (tree *tp, walk_tree_fn func, void *data,
 	case OMP_CLAUSE_GANG:
 	case OMP_CLAUSE_ASYNC:
 	case OMP_CLAUSE_WAIT:
-	case OMP_NO_CLAUSE_CACHE:
 	case OMP_CLAUSE_WORKER:
 	case OMP_CLAUSE_VECTOR:
 	case OMP_CLAUSE_NUM_GANGS:
@@ -11194,6 +11193,7 @@ walk_tree_1 (tree *tp, walk_tree_fn func, void *data,
 	case OMP_CLAUSE_FROM:
 	case OMP_CLAUSE_TO:
 	case OMP_CLAUSE_MAP:
+	case OMP_CLAUSE__CACHE_:
 	  WALK_SUBTREE (OMP_CLAUSE_DECL (*tp));
 	  WALK_SUBTREE (OMP_CLAUSE_OPERAND (*tp, 1));
 	  WALK_SUBTREE_TAIL (OMP_CLAUSE_CHAIN (*tp));
diff --git gcc/tree.def gcc/tree.def
index 871a7fb..f44853a 100644
--- gcc/tree.def
+++ gcc/tree.def
@@ -1163,8 +1163,9 @@ DEFTREECODE (OACC_ENTER_DATA, "oacc_enter_data", tcc_statement, 1)
    Operand 0: OACC_EXIT_DATA_CLAUSES: List of clauses.  */
 DEFTREECODE (OACC_EXIT_DATA, "oacc_exit_data", tcc_statement, 1)
 
-/* OpenACC - #pragma acc cache [clause1 ... clauseN]
-   Operand 0: OACC_CACHE_CLAUSES: List of clauses.  */
+/* OpenACC - #pragma acc cache (variable1 ... variableN)
+   Operand 0: OACC_CACHE_CLAUSES: List of variables (transformed into
+	OMP_CLAUSE__CACHE_ clauses).  */
 DEFTREECODE (OACC_CACHE, "oacc_cache", tcc_statement, 1)
 
 /* OpenMP - #pragma omp target update [clause1 ... clauseN]
diff --git gcc/tree.h gcc/tree.h
index c91e716..e1adbab 100644
--- gcc/tree.h
+++ gcc/tree.h
@@ -1254,7 +1254,7 @@ extern void protected_set_expr_location (tree, location_t);
 #define OMP_CLAUSE_SIZE(NODE)						\
   OMP_CLAUSE_OPERAND (OMP_CLAUSE_RANGE_CHECK (OMP_CLAUSE_CHECK (NODE),	\
 					      OMP_CLAUSE_FROM,		\
-					      OMP_CLAUSE_MAP), 1)
+					      OMP_CLAUSE__CACHE_), 1)
 
 #define OMP_CLAUSE_CHAIN(NODE)     TREE_CHAIN (OMP_CLAUSE_CHECK (NODE))
 #define OMP_CLAUSE_DECL(NODE)      					\


Grüße,
 Thomas

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