From fa6894ec9ce25f5aff10ec176212383f5c88b1ec Mon Sep 17 00:00:00 2001 From: Marcel Vollweiler Date: Wed, 12 May 2021 09:31:58 -0700 Subject: [PATCH] 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. --- gcc/c/c-parser.c | 120 ++++++++++----------- gcc/cp/parser.c | 90 ++++++++++++---- gcc/testsuite/c-c++-common/gomp/map-6.c | 135 ++++++++++++++++++++++++ gcc/testsuite/c-c++-common/gomp/map-7.c | 20 ++++ 4 files changed, 285 insertions(+), 80 deletions(-) create mode 100644 gcc/testsuite/c-c++-common/gomp/map-6.c create mode 100644 gcc/testsuite/c-c++-common/gomp/map-7.c diff --git a/gcc/c/c-parser.c b/gcc/c/c-parser.c index 5cdeb21a458..f79b8395366 100644 --- 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; - } - else if (sectok->type == CPP_NAME) - { - 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 % 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 % modifiers"); + parens.skip_until_found_close (parser); + return list; + } + close_modifier++; + } + else + { + c_parser_error (parser, "%<#pragma omp target%> with " + "modifier other than % or %" + "on % 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 0fe29c658d2..41df5dd525f 100644 --- 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 % 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 % 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 % or %" + "on % 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 index 00000000000..6ee59714847 --- /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 index 00000000000..3f1e9729e41 --- /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) + ; +}