OpenMP: Add support for 'close' in map clause

gcc/c/ChangeLog:

	* c-parser.c (c_parser_omp_clause_map): Support map-type-modifier
	'close'.

gcc/cp/ChangeLog:

	* parser.c (cp_parser_omp_clause_map): Support map-type-modifier
	'close'.

gcc/testsuite/ChangeLog:

	* c-c++-common/gomp/map-6.c: New test.
	* c-c++-common/gomp/map-7.c: New test.

diff --git a/gcc/c/c-parser.c b/gcc/c/c-parser.c
index 5cdeb21..f79b839 100644 (file)
--- a/gcc/c/c-parser.c
+++ b/gcc/c/c-parser.c
@@ -15643,54 +15643,83 @@ c_parser_omp_clause_depend (c_parser *parser, tree list)
    map-kind:
      alloc | to | from | tofrom | release | delete
 
-   map ( always [,] map-kind: variable-list ) */
+   map ( always [,] map-kind: variable-list )
+
+   OpenMP 5.0:
+   map ( [map-type-modifier[,] ...] map-kind: variable-list )
+
+   map-type-modifier:
+     always | close */
 
 static tree
 c_parser_omp_clause_map (c_parser *parser, tree list)
 {
   location_t clause_loc = c_parser_peek_token (parser)->location;
   enum gomp_map_kind kind = GOMP_MAP_TOFROM;
-  int always = 0;
-  enum c_id_kind always_id_kind = C_ID_NONE;
-  location_t always_loc = UNKNOWN_LOCATION;
-  tree always_id = NULL_TREE;
   tree nl, c;
 
   matching_parens parens;
   if (!parens.require_open (parser))
     return list;
 
-  if (c_parser_next_token_is (parser, CPP_NAME))
+  int pos = 1;
+  int map_kind_pos = 0;
+  while (c_parser_peek_nth_token_raw (parser, pos)->type == CPP_NAME)
+    {
+      if (c_parser_peek_nth_token_raw (parser, pos + 1)->type == CPP_COLON)
+       {
+         map_kind_pos = pos;
+         break;
+       }
+
+      if (c_parser_peek_nth_token_raw (parser, pos + 1)->type == CPP_COMMA)
+       pos++;
+      pos++;
+    }
+
+  int always_modifier = 0;
+  int close_modifier = 0;
+  for (int pos = 1; pos < map_kind_pos; ++pos)
     {
       c_token *tok = c_parser_peek_token (parser);
+
+      if (tok->type == CPP_COMMA)
+       {
+         c_parser_consume_token (parser);
+         continue;
+       }
+
       const char *p = IDENTIFIER_POINTER (tok->value);
-      always_id_kind = tok->id_kind;
-      always_loc = tok->location;
-      always_id = tok->value;
       if (strcmp ("always", p) == 0)
        {
-         c_token *sectok = c_parser_peek_2nd_token (parser);
-         if (sectok->type == CPP_COMMA)
+         if (always_modifier)
            {
-             c_parser_consume_token (parser);
-             c_parser_consume_token (parser);
-             always = 2;
+             c_parser_error (parser, "too many %<always%> modifiers");
+             parens.skip_until_found_close (parser);
+             return list;
            }
-         else if (sectok->type == CPP_NAME)
+         always_modifier++;
+       }
+      else if (strcmp ("close", p) == 0)
+       {
+         if (close_modifier)
            {
-             p = IDENTIFIER_POINTER (sectok->value);
-             if (strcmp ("alloc", p) == 0
-                 || strcmp ("to", p) == 0
-                 || strcmp ("from", p) == 0
-                 || strcmp ("tofrom", p) == 0
-                 || strcmp ("release", p) == 0
-                 || strcmp ("delete", p) == 0)
-               {
-                 c_parser_consume_token (parser);
-                 always = 1;
-               }
+             c_parser_error (parser, "too many %<close%> modifiers");
+             parens.skip_until_found_close (parser);
+             return list;
            }
+         close_modifier++;
        }
+      else
+       {
+         c_parser_error (parser, "%<#pragma omp target%> with "
+                                 "modifier other than %<always%> or %<close%>"
+                                 "on %<map%> clause");
+         parens.skip_until_found_close (parser);
+         return list;
+       }
+
+       c_parser_consume_token (parser);
     }
 
   if (c_parser_next_token_is (parser, CPP_NAME)
@@ -15700,11 +15729,11 @@ c_parser_omp_clause_map (c_parser *parser, tree list)
       if (strcmp ("alloc", p) == 0)
        kind = GOMP_MAP_ALLOC;
       else if (strcmp ("to", p) == 0)
-       kind = always ? GOMP_MAP_ALWAYS_TO : GOMP_MAP_TO;
+       kind = always_modifier ? GOMP_MAP_ALWAYS_TO : GOMP_MAP_TO;
       else if (strcmp ("from", p) == 0)
-       kind = always ? GOMP_MAP_ALWAYS_FROM : GOMP_MAP_FROM;
+       kind = always_modifier ? GOMP_MAP_ALWAYS_FROM : GOMP_MAP_FROM;
       else if (strcmp ("tofrom", p) == 0)
-       kind = always ? GOMP_MAP_ALWAYS_TOFROM : GOMP_MAP_TOFROM;
+       kind = always_modifier ? GOMP_MAP_ALWAYS_TOFROM : GOMP_MAP_TOFROM;
       else if (strcmp ("release", p) == 0)
        kind = GOMP_MAP_RELEASE;
       else if (strcmp ("delete", p) == 0)
@@ -15719,35 +15748,6 @@ c_parser_omp_clause_map (c_parser *parser, tree list)
       c_parser_consume_token (parser);
       c_parser_consume_token (parser);
     }
-  else if (always)
-    {
-      if (always_id_kind != C_ID_ID)
-       {
-         c_parser_error (parser, "expected identifier");
-         parens.skip_until_found_close (parser);
-         return list;
-       }
-
-      tree t = lookup_name (always_id);
-      if (t == NULL_TREE)
-       {
-         undeclared_variable (always_loc, always_id);
-         t = error_mark_node;
-       }
-      if (t != error_mark_node)
-       {
-         tree u = build_omp_clause (clause_loc, OMP_CLAUSE_MAP);
-         OMP_CLAUSE_DECL (u) = t;
-         OMP_CLAUSE_CHAIN (u) = list;
-         OMP_CLAUSE_SET_MAP_KIND (u, kind);
-         list = u;
-       }
-      if (always == 1)
-       {
-         parens.skip_until_found_close (parser);
-         return list;
-       }
-    }
 
   nl = c_parser_omp_variable_list (parser, clause_loc, OMP_CLAUSE_MAP, list);
 

diff --git a/gcc/cp/parser.c b/gcc/cp/parser.c
index 0fe29c6..41df5dd 100644 (file)
--- a/gcc/cp/parser.c
+++ b/gcc/cp/parser.c
@@ -37859,40 +37859,90 @@ cp_parser_omp_clause_depend (cp_parser *parser, tree list, location_t loc)
    map-kind:
      alloc | to | from | tofrom | release | delete
 
-   map ( always [,] map-kind: variable-list ) */
+   map ( always [,] map-kind: variable-list )
+
+   OpenMP 5.0:
+   map ( [map-type-modifier[,] ...] map-kind: variable-list )
+
+   map-type-modifier:
+     always | close */
 
 static tree
 cp_parser_omp_clause_map (cp_parser *parser, tree list)
 {
   tree nlist, c;
   enum gomp_map_kind kind = GOMP_MAP_TOFROM;
-  bool always = false;
 
   if (!cp_parser_require (parser, CPP_OPEN_PAREN, RT_OPEN_PAREN))
     return list;
 
-  if (cp_lexer_next_token_is (parser->lexer, CPP_NAME))
+  int pos = 1;
+  int map_kind_pos = 0;
+  while (cp_lexer_peek_nth_token (parser->lexer, pos)->type == CPP_NAME
+        || cp_lexer_peek_nth_token (parser->lexer, pos)->keyword == RID_DELETE)
     {
-      tree id = cp_lexer_peek_token (parser->lexer)->u.value;
-      const char *p = IDENTIFIER_POINTER (id);
+      if (cp_lexer_peek_nth_token (parser->lexer, pos + 1)->type == CPP_COLON)
+       {
+         map_kind_pos = pos;
+         break;
+       }
+
+      if (cp_lexer_peek_nth_token (parser->lexer, pos + 1)->type == CPP_COMMA)
+       pos++;
+      pos++;
+    }
 
+  bool always_modifier = false;
+  bool close_modifier = false;
+  for (int pos = 1; pos < map_kind_pos; ++pos)
+    {
+      cp_token *tok = cp_lexer_peek_token (parser->lexer);
+      if (tok->type == CPP_COMMA)
+       {
+         cp_lexer_consume_token (parser->lexer);
+         continue;
+       }
+
+      const char *p = IDENTIFIER_POINTER (tok->u.value);
       if (strcmp ("always", p) == 0)
        {
-         int nth = 2;
-         if (cp_lexer_peek_nth_token (parser->lexer, 2)->type == CPP_COMMA)
-           nth++;
-         if ((cp_lexer_peek_nth_token (parser->lexer, nth)->type == CPP_NAME
-              || (cp_lexer_peek_nth_token (parser->lexer, nth)->keyword
-                  == RID_DELETE))
-             && (cp_lexer_peek_nth_token (parser->lexer, nth + 1)->type
-                 == CPP_COLON))
+         if (always_modifier)
            {
-             always = true;
-             cp_lexer_consume_token (parser->lexer);
-             if (nth == 3)
-               cp_lexer_consume_token (parser->lexer);
+             cp_parser_error (parser, "too many %<always%> modifiers");
+             cp_parser_skip_to_closing_parenthesis (parser,
+                                                    /*recovering=*/true,
+                                                    /*or_comma=*/false,
+                                                    /*consume_paren=*/true);
+             return list;
+           }
+         always_modifier = true;
+       }
+      else if (strcmp ("close", p) == 0)
+       {
+         if (close_modifier)
+           {
+             cp_parser_error (parser, "too many %<close%> modifiers");
+             cp_parser_skip_to_closing_parenthesis (parser,
+                                                    /*recovering=*/true,
+                                                    /*or_comma=*/false,
+                                                    /*consume_paren=*/true);
+             return list;
            }
+         close_modifier = true;
        }
+      else
+       {
+         cp_parser_error (parser, "%<#pragma omp target%> with "
+                                  "modifier other than %<always%> or %<close%>"
+                                  "on %<map%> clause");
+         cp_parser_skip_to_closing_parenthesis (parser,
+                                                /*recovering=*/true,
+                                                /*or_comma=*/false,
+                                                /*consume_paren=*/true);
+         return list;
+       }
+
+       cp_lexer_consume_token (parser->lexer);
     }
 
   if (cp_lexer_next_token_is (parser->lexer, CPP_NAME)
@@ -37904,11 +37954,11 @@ cp_parser_omp_clause_map (cp_parser *parser, tree list)
       if (strcmp ("alloc", p) == 0)
        kind = GOMP_MAP_ALLOC;
       else if (strcmp ("to", p) == 0)
-       kind = always ? GOMP_MAP_ALWAYS_TO : GOMP_MAP_TO;
+       kind = always_modifier ? GOMP_MAP_ALWAYS_TO : GOMP_MAP_TO;
       else if (strcmp ("from", p) == 0)
-       kind = always ? GOMP_MAP_ALWAYS_FROM : GOMP_MAP_FROM;
+       kind = always_modifier ? GOMP_MAP_ALWAYS_FROM : GOMP_MAP_FROM;
       else if (strcmp ("tofrom", p) == 0)
-       kind = always ? GOMP_MAP_ALWAYS_TOFROM : GOMP_MAP_TOFROM;
+       kind = always_modifier ? GOMP_MAP_ALWAYS_TOFROM : GOMP_MAP_TOFROM;
       else if (strcmp ("release", p) == 0)
        kind = GOMP_MAP_RELEASE;
       else

diff --git a/gcc/testsuite/c-c++-common/gomp/map-6.c b/gcc/testsuite/c-c++-common/gomp/map-6.c
new file mode 100644 (file)
index 0000000..6ee5971
--- /dev/null
+++ b/gcc/testsuite/c-c++-common/gomp/map-6.c
@@ -0,0 +1,135 @@
+/* { dg-do compile } */
+/* { dg-additional-options "-fdump-tree-original" } */
+
+void
+foo (void)
+{
+  /* Test to ensure that the close modifier is parsed and ignored in map clauses. */
+  int a, b, b1, b2, b3, b4, b5, b6, b7;
+
+  #pragma omp target map (a)
+  ;
+
+  #pragma omp target map (to:a)
+  ;
+
+  #pragma omp target map (a to: b) /* { dg-error "'#pragma omp target' with modifier other than 'always' or 'close'" } */
+  ;
+
+  #pragma omp target map (close, a to: b) /* { dg-error "'#pragma omp target' with modifier other than 'always' or 'close'" } */
+  ;
+
+  #pragma omp target map (close a) /* { dg-error "'close' undeclared" "" { target c } } */ 
+  /* { dg-error "'close' has not been declared" "" { target c++ } .-1 } */ 
+  /* { dg-error "expected '\\)' before 'a'" "" { target *-*-* } .-2 } */
+  ;
+
+  #pragma omp target map (always a) /* { dg-error "'always' undeclared" "" { target c } } */
+  /* { dg-error "'always' has not been declared" "" { target c++ } .-1 } */ 
+  /* { dg-error "expected '\\)' before 'a'" "" { target *-*-* } .-2 } */
+  ;
+
+  #pragma omp target map (close to:a)
+  ;
+
+  #pragma omp target map (close, to:a)
+  ;
+
+  #pragma omp target map (close delete:a) /* { dg-error "'#pragma omp target' with map-type other than 'to', 'from', 'tofrom' or 'alloc' on 'map' clause" } */
+  ;
+
+  #pragma omp target map (close always to:b1)
+  ;
+
+  #pragma omp target map (close, always to:b2)
+  ;
+
+  #pragma omp target map (close, always, to:b3)
+  ;
+
+  #pragma omp target map (always close to:b4)
+  ;
+
+  #pragma omp target map (always, close to:b5)
+  ;
+
+  #pragma omp target map (always, close, to:b6)
+  ;
+
+  #pragma omp target map (always, always, to:a) /* { dg-error "too many 'always' modifiers" } */
+  ;
+
+  #pragma omp target map (always always, to:a) /* { dg-error "too many 'always' modifiers" } */
+  ;
+
+  #pragma omp target map (always, always to:a) /* { dg-error "too many 'always' modifiers" } */
+  ;
+
+  #pragma omp target map (always always to:a) /* { dg-error "too many 'always' modifiers" } */
+  ;
+
+  #pragma omp target map (close, close, to:a) /* { dg-error "too many 'close' modifiers" } */
+  ;
+
+  #pragma omp target map (close close, to:a) /* { dg-error "too many 'close' modifiers" } */
+  ;
+
+  #pragma omp target map (close, close to:a) /* { dg-error "too many 'close' modifiers" } */
+  ;
+
+  #pragma omp target map (close close to:a) /* { dg-error "too many 'close' modifiers" } */
+  ;
+
+  #pragma omp target map (always to : a) map (close to : b)
+  ;
+
+  int close = 0;
+  #pragma omp target map (close) 
+  ;
+
+  #pragma omp target map (close a) /* { dg-error "expected '\\)' before 'a'" } */ 
+  ;
+
+  int always = 0;
+  #pragma omp target map (always)
+  ;
+
+  #pragma omp target map (always a) /* { dg-error "expected '\\)' before 'a'" } */
+  ;
+
+  #pragma omp target map (always, close)
+  ;
+
+  #pragma omp target map (always, always)  /* { dg-error "'always' appears more than once in map clauses" } */
+  ;
+
+  #pragma omp target map (always, always, close)  /* { dg-error "'always' appears more than once in map clauses" } */
+  ;
+
+  #pragma omp target map (always, close, to: always, close, b7)
+  ;
+
+  int to = 0;
+  #pragma omp target map (always, close, to)
+  ;
+
+  #pragma omp target map (to, always, close)
+    {
+      to = always = close = 1;
+    }
+  if (to != 1 || always != 1 || close != 1)
+    __builtin_abort ();
+  ;
+}
+
+/* { dg-final { scan-tree-dump-not "map\\(\[^\n\r)]*close\[^\n\r)]*to:" "original" } } */
+
+/* { dg-final { scan-tree-dump-times "pragma omp target map\\(always,to:" 7 "original" } } */
+
+/* { dg-final { scan-tree-dump "pragma omp target map\\(always,to:b1" "original" } } */
+/* { dg-final { scan-tree-dump "pragma omp target map\\(always,to:b2" "original" } } */
+/* { dg-final { scan-tree-dump "pragma omp target map\\(always,to:b3" "original" } } */
+/* { dg-final { scan-tree-dump "pragma omp target map\\(always,to:b4" "original" } } */
+/* { dg-final { scan-tree-dump "pragma omp target map\\(always,to:b5" "original" } } */
+/* { dg-final { scan-tree-dump "pragma omp target map\\(always,to:b6" "original" } } */
+/* { dg-final { scan-tree-dump "pragma omp target map\\(always,to:b7\\) map\\(always,to:close\\) map\\(always,to:always\\)" "original" } } */

diff --git a/gcc/testsuite/c-c++-common/gomp/map-7.c b/gcc/testsuite/c-c++-common/gomp/map-7.c
new file mode 100644 (file)
index 0000000..3f1e972
--- /dev/null
+++ b/gcc/testsuite/c-c++-common/gomp/map-7.c
@@ -0,0 +1,20 @@
+/* { dg-do compile } */
+
+void
+foo (void)
+{
+  /* Test to ensure that the close modifier is parsed and ignored in map clauses. */
+
+  #define N 1024
+  int always[N];
+  int close;
+
+  #pragma omp target map(always[:N]) 
+  ;
+
+  #pragma omp target map(close, always[:N]) 
+  ;
+
+  #pragma omp target map(always[:N], close) 
+  ;
+}