Am 10.05.2021 um 20:34 schrieb Jakub Jelinek:
On Mon, May 10, 2021 at 04:11:39PM +0200, Marcel Vollweiler wrote:
@@ -15660,37 +15665,54 @@ c_parser_omp_clause_map (c_parser *parser, tree list)
if (!parens.require_open (parser))
return list;
- if (c_parser_next_token_is (parser, CPP_NAME))
+ int always = 0;
+ int close = 0;
+ int pos = 1;
+ while (c_parser_peek_nth_token_raw (parser, pos)->type == CPP_NAME)
Nice, totally missed that Joseph has added this.
{
- c_token *tok = c_parser_peek_token (parser);
+ c_token *tok = c_parser_peek_nth_token_raw (parser, pos);
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)
{
- c_parser_consume_token (parser);
- c_parser_consume_token (parser);
- always = 2;
+ c_parser_error (parser, "expected modifier %<always%> only once");
The usual wording would be
"too many %<always%> modifiers"
Changed for 'always' and 'close' for C and C++.
+ parens.skip_until_found_close (parser);
+ return list;
+ }
+
+ always_id_kind = tok->id_kind;
+ always_loc = tok->location;
+ always_id = tok->value;
But you don't need any of the always_{id_kind,loc,id} variables anymore,
so they should be removed and everything that touches them too.
That's true. I removed them.
+
+ always++;
+ }
+ else if (strcmp ("close", p) == 0)
+ {
+ if (close)
+ {
+ c_parser_error (parser, "expected modifier %<close%> only once");
Similarly.
+ parens.skip_until_found_close (parser);
+ return list;
}
- else if (sectok->type == CPP_NAME)
+
+ close++;
+ }
+ else if (c_parser_peek_nth_token_raw (parser, pos + 1)->type ==
CPP_COLON)
IMHO you should at least check that tok->type == CPP_NAME before
checking pos + 1 token's type, you don't want to skip over CPP_EOF,
CPP_PRAGMA_EOF, or even CPP_CLOSE_PAREN etc.
Perhaps by adding
if (tok->type != CPP_NAME)
break;
right after c_token *tok = c_parser_peek_nth_token_raw (parser, pos); ?
The check of the token's type at position 'pos' is done in the condition
of the while loop, that means
'c_parser_peek_nth_token_raw (parser, pos + 1)->type == CPP_COLON'
is only reached when
'c_parser_peek_nth_token_raw (parser, pos)->type == CPP_NAME'
holds (since 'pos' is not changed in between).
+ {
+ for (int i = 1; i < pos; ++i)
{
- 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_peek_token(parser);
Formatting, space before (
Corrected.
+ c_parser_consume_token (parser);
}
+ break;
And, IMHO something should clear always and close (btw, might be better
to use close_modifier as variable name and for consistency always_modifier)
unless we reach the CPP_COLON case.
Good point, I agree with both. Cleared and renamed :)
Because we don't want
map (always, close)
to imply
map (always, close, tofrom: always, close)
but
map (tofrom: always, close)
and my reading of your changes suggests that we actually use the
*_ALWAYS* kinds in that case.
+ cp_parser_error (parser,
+ "expected modifier %<always%> only once");
See above.
+ cp_parser_skip_to_closing_parenthesis (parser,
+ /*recovering=*/true,
+ /*or_comma=*/false,
+ /*consume_paren=*/true);
+ return list;
+ }
+
+ always = true;
+ }
+ else if (strcmp ("close", p) == 0)
+ {
+ if (close)
+ {
+ cp_parser_error (parser,
+ "expected modifier %<close%> only once");
Likewise.
+ else if (cp_lexer_peek_nth_token (parser->lexer, pos + 1)->type
+ == CPP_COLON)
+ {
+ for (int i = 1; i < pos; ++i)
+ cp_lexer_consume_token (parser->lexer);
+ break;
+ }
+ else
+ break;
+
+ if (cp_lexer_peek_nth_token (parser->lexer, pos + 1)->type == CPP_COMMA)
+ pos++;
+ pos++;
}
Again, I don't see anything that would clear always/close if it didn't reach
the CPP_COLON case.
And it should be covered in the testcase.
I added a test case similar to your example above:
'#pragma omp target map (always, close, to: always, close, b7)'
Jakub
The new version of the patch including the changes is attached.
Thanks,
Marcel
-----------------
Mentor Graphics (Deutschland) GmbH, Arnulfstrasse 201, 80634 München
Registergericht München HRB 106955, Geschäftsführer: Thomas Heurung, Frank
Thürauf
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..310df89 100644
--- a/gcc/c/c-parser.c
+++ b/gcc/c/c-parser.c
@@ -15643,54 +15643,73 @@ 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 always_modifier = 0;
+ int close_modifier = 0;
+ int pos = 1;
+ while (c_parser_peek_nth_token_raw (parser, pos)->type == CPP_NAME)
{
- c_token *tok = c_parser_peek_token (parser);
+ c_token *tok = c_parser_peek_nth_token_raw (parser, pos);
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;
+ }
+
+ always_modifier++;
+ }
+ else if (strcmp ("close", p) == 0)
+ {
+ if (close_modifier)
+ {
+ c_parser_error (parser, "too many %<close%> modifiers");
+ parens.skip_until_found_close (parser);
+ return list;
}
- else if (sectok->type == CPP_NAME)
+
+ close_modifier++;
+ }
+ else if (c_parser_peek_nth_token_raw (parser, pos + 1)->type ==
CPP_COLON)
+ {
+ for (int i = 1; i < pos; ++i)
{
- 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_peek_token (parser);
+ c_parser_consume_token (parser);
}
+ break;
}
+ else
+ {
+ always_modifier = 0;
+ close_modifier = 0;
+ break;
+ }
+
+ if (c_parser_peek_nth_token_raw (parser, pos + 1)->type == CPP_COMMA)
+ pos++;
+ pos++;
}
if (c_parser_next_token_is (parser, CPP_NAME)
@@ -15700,11 +15719,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 +15738,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 99eccf0..55a59ed 100644
--- a/gcc/cp/parser.c
+++ b/gcc/cp/parser.c
@@ -37840,40 +37840,77 @@ 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;
+ bool always_modifier = false;
+ bool close_modifier = false;
+ int pos = 1;
if (!cp_parser_require (parser, CPP_OPEN_PAREN, RT_OPEN_PAREN))
return list;
- if (cp_lexer_next_token_is (parser->lexer, CPP_NAME))
+ 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;
+ tree id = cp_lexer_peek_nth_token (parser->lexer, pos)->u.value;
const char *p = IDENTIFIER_POINTER (id);
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 if (cp_lexer_peek_nth_token (parser->lexer, pos + 1)->type
+ == CPP_COLON)
+ {
+ for (int i = 1; i < pos; ++i)
+ cp_lexer_consume_token (parser->lexer);
+ break;
+ }
+ else
+ {
+ always_modifier = 0;
+ close_modifier = 0;
+ break;
+ }
+
+ if (cp_lexer_peek_nth_token (parser->lexer, pos + 1)->type == CPP_COMMA)
+ pos++;
+ pos++;
}
if (cp_lexer_next_token_is (parser->lexer, CPP_NAME)
@@ -37885,11 +37922,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
index 0000000..b524e30
--- /dev/null
+++ b/gcc/testsuite/c-c++-common/gomp/map-6.c
@@ -0,0 +1,126 @@
+/* { 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 "expected '\\)' before 'to'"
} */
+ ;
+
+ #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, 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
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)
+ ;
+}
