This patch adds device-modifiers to the device clause:
#pragma omp target device ([ device-modifier :] integer-expression)
where device-modifier is either 'ancestor' or 'device_num'.
The 'device_num' case
#pragma omp target device (device_num : integer-expression)
is treated in the same way as
#pragma omp target device (integer-expression)
before.
For the 'ancestor' case
#pragma omp target device (ancestor: integer-expression)
a message 'sorry, not yet implemented' is output.
-----------------
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 device-modifiers for 'omp target device'
gcc/c/ChangeLog:
* c-parser.c (c_parser_omp_clause_device): Add support for
device-modifiers for 'omp target device'.
gcc/cp/ChangeLog:
* parser.c (cp_parser_omp_clause_device): Add support for
device-modifiers for 'omp target device'.
gcc/fortran/ChangeLog:
* openmp.c (gfc_match_omp_clauses): Add support for
device-modifiers for 'omp target device'.
gcc/testsuite/ChangeLog:
* c-c++-common/gomp/target-device-1.c: New test.
* c-c++-common/gomp/target-device-2.c: New test.
* gfortran.dg/gomp/target-device-1.f90: New test.
* gfortran.dg/gomp/target-device-2.f90: New test.
diff --git a/gcc/c/c-parser.c b/gcc/c/c-parser.c
index 9a56e0c..defc52d 100644
--- a/gcc/c/c-parser.c
+++ b/gcc/c/c-parser.c
@@ -15864,37 +15864,117 @@ c_parser_omp_clause_map (c_parser *parser, tree list)
}
/* OpenMP 4.0:
- device ( expression ) */
+ device ( expression )
+
+ OpenMP 5.0:
+ device ( [device-modifier :] integer-expression )
+
+ device-modifier:
+ ancestor | device_num */
static tree
c_parser_omp_clause_device (c_parser *parser, tree list)
{
location_t clause_loc = c_parser_peek_token (parser)->location;
+ location_t expr_loc;
+ c_expr expr;
+ tree c, t;
+
matching_parens parens;
- if (parens.require_open (parser))
+ if (!parens.require_open (parser))
+ return list;
+
+ int pos = 1;
+ int pos_colon = 0;
+ while (c_parser_peek_nth_token_raw (parser, pos)->type == CPP_NAME
+ || c_parser_peek_nth_token_raw (parser, pos)->type == CPP_COLON
+ || c_parser_peek_nth_token_raw (parser, pos)->type == CPP_COMMA)
{
- location_t expr_loc = c_parser_peek_token (parser)->location;
- c_expr expr = c_parser_expr_no_commas (parser, NULL);
- expr = convert_lvalue_to_rvalue (expr_loc, expr, false, true);
- tree c, t = expr.value;
- t = c_fully_fold (t, false, NULL);
+ if (c_parser_peek_nth_token_raw (parser, pos)->type == CPP_COLON)
+ {
+ pos_colon = pos;
+ break;
+ }
+ pos++;
+ }
- parens.skip_until_found_close (parser);
+ const char *err_msg;
+ if (pos_colon == 1)
+ {
+ err_msg = "expected device-modifier %<ancestor%> or %<device_num%>";
+ goto invalid_kind;
+ }
- if (!INTEGRAL_TYPE_P (TREE_TYPE (t)))
+ if (pos_colon > 1)
+ {
+ if (c_parser_peek_nth_token_raw (parser, 1)->type == CPP_NAME)
{
- c_parser_error (parser, "expected integer expression");
- return list;
+ c_token *tok = c_parser_peek_token (parser);
+ const char *p = IDENTIFIER_POINTER (tok->value);
+ if (strcmp ("ancestor", p) == 0)
+ {
+ if (pos_colon > 2)
+ {
+ err_msg = "expected only one device-modifier %<ancestor%> or "
+ "%<device_num%>";
+ goto invalid_kind;
+ }
+
+ sorry_at (tok->location, "%<ancestor%> not yet supported");
+ c_parser_skip_until_found (parser, CPP_CLOSE_PAREN, NULL);
+ return list;
+ }
+ else if (strcmp ("device_num", p) == 0)
+ {
+ if (pos_colon > 2)
+ {
+ err_msg = "expected only one device-modifier %<ancestor%> or "
+ "%<device_num%>";
+ goto invalid_kind;
+ }
+ c_parser_consume_token (parser);
+ c_parser_peek_token (parser);
+ c_parser_consume_token (parser);
+ }
+ else
+ {
+ err_msg = "expected device-modifier %<ancestor%> or "
+ "%<device_num%>";
+ goto invalid_kind;
+ }
+ }
+ else
+ {
+ err_msg = "expected device-modifier %<ancestor%> or %<device_num%>";
+ goto invalid_kind;
}
+ }
- check_no_duplicate_clause (list, OMP_CLAUSE_DEVICE, "device");
+ expr_loc = c_parser_peek_token (parser)->location;
+ expr = c_parser_expr_no_commas (parser, NULL);
+ expr = convert_lvalue_to_rvalue (expr_loc, expr, false, true);
+ c, t = expr.value;
+ t = c_fully_fold (t, false, NULL);
- c = build_omp_clause (clause_loc, OMP_CLAUSE_DEVICE);
- OMP_CLAUSE_DEVICE_ID (c) = t;
- OMP_CLAUSE_CHAIN (c) = list;
- list = c;
+ parens.skip_until_found_close (parser);
+
+ if (!INTEGRAL_TYPE_P (TREE_TYPE (t)))
+ {
+ c_parser_error (parser, "expected integer expression");
+ return list;
}
+ check_no_duplicate_clause (list, OMP_CLAUSE_DEVICE, "device");
+
+ c = build_omp_clause (clause_loc, OMP_CLAUSE_DEVICE);
+ OMP_CLAUSE_DEVICE_ID (c) = t;
+ OMP_CLAUSE_CHAIN (c) = list;
+ list = c;
+ return list;
+
+ invalid_kind:
+ c_parser_error (parser, err_msg);
+ parens.skip_until_found_close (parser);
return list;
}
diff --git a/gcc/cp/parser.c b/gcc/cp/parser.c
index 93698aa..9c7dfa7 100644
--- a/gcc/cp/parser.c
+++ b/gcc/cp/parser.c
@@ -38536,7 +38536,13 @@ cp_parser_omp_clause_map (cp_parser *parser, tree list)
}
/* OpenMP 4.0:
- device ( expression ) */
+ device ( expression )
+
+ OpenMP 5.0:
+ device ( [device-modifier :] integer-expression )
+
+ device-modifier:
+ ancestor | device_num */
static tree
cp_parser_omp_clause_device (cp_parser *parser, tree list,
@@ -38548,6 +38554,75 @@ cp_parser_omp_clause_device (cp_parser *parser, tree
list,
if (!parens.require_open (parser))
return list;
+ int pos = 1;
+ int pos_colon = 0;
+ while (cp_lexer_peek_nth_token (parser->lexer, pos)->type == CPP_NAME
+ || cp_lexer_peek_nth_token (parser->lexer, pos)->type == CPP_COLON
+ || cp_lexer_peek_nth_token (parser->lexer, pos)->type == CPP_COMMA)
+ {
+ if (cp_lexer_peek_nth_token (parser->lexer, pos)->type == CPP_COLON)
+ {
+ pos_colon = pos;
+ break;
+ }
+ pos++;
+ }
+
+ const char *err_msg;
+ if (pos_colon == 1)
+ {
+ err_msg = "expected device-modifier %<ancestor%> or %<device_num%>";
+ goto invalid_kind;
+ }
+
+ if (pos_colon > 1)
+ {
+ if (cp_lexer_peek_nth_token (parser->lexer, 1)->type == CPP_NAME)
+ {
+ cp_token *tok = cp_lexer_peek_token (parser->lexer);
+ const char *p = IDENTIFIER_POINTER (tok->u.value);
+ if (strcmp ("ancestor", p) == 0)
+ {
+ if (pos_colon > 2)
+ {
+ err_msg = "expected only one device-modifier %<ancestor%> or "
+ "%<device_num%>";
+ goto invalid_kind;
+ }
+
+ sorry_at (tok->location, "%<ancestor%> not yet supported");
+ cp_parser_skip_to_closing_parenthesis (parser,
+ /*recovering=*/true,
+ /*or_comma=*/false,
+ /*consume_paren=*/true);
+ return list;
+ }
+ else if (strcmp ("device_num", p) == 0)
+ {
+ if (pos_colon > 2)
+ {
+ err_msg = "expected only one device-modifier %<ancestor%> or "
+ "%<device_num%>";
+ goto invalid_kind;
+ }
+ cp_lexer_consume_token (parser->lexer);
+ cp_lexer_peek_token (parser->lexer);
+ cp_lexer_consume_token (parser->lexer);
+ }
+ else
+ {
+ err_msg = "expected device-modifier %<ancestor%> or "
+ "%<device_num%>";
+ goto invalid_kind;
+ }
+ }
+ else
+ {
+ err_msg = "expected device-modifier %<ancestor%> or %<device_num%>";
+ goto invalid_kind;
+ }
+ }
+
t = cp_parser_assignment_expression (parser);
if (t == error_mark_node
@@ -38564,6 +38639,14 @@ cp_parser_omp_clause_device (cp_parser *parser, tree
list,
OMP_CLAUSE_CHAIN (c) = list;
return c;
+
+ invalid_kind:
+ cp_parser_error (parser, err_msg);
+ cp_parser_skip_to_closing_parenthesis (parser,
+ /*recovering=*/true,
+ /*or_comma=*/false,
+ /*consume_paren=*/true);
+ return list;
}
/* OpenMP 4.0:
diff --git a/gcc/fortran/openmp.c b/gcc/fortran/openmp.c
index 357a1e1..ac2e18a 100644
--- a/gcc/fortran/openmp.c
+++ b/gcc/fortran/openmp.c
@@ -1714,8 +1714,33 @@ gfc_match_omp_clauses (gfc_omp_clauses **cp, const
omp_mask mask,
if ((mask & OMP_CLAUSE_DEVICE)
&& !openacc
&& c->device == NULL
- && gfc_match ("device ( %e )", &c->device) == MATCH_YES)
- continue;
+ && gfc_match ("device ( ") == MATCH_YES)
+ {
+ if (gfc_match ("device_num : ") == MATCH_YES)
+ {
+ if (gfc_match ("%e )", &c->device) != MATCH_YES)
+ {
+ gfc_error ("Expected integer expression at %C");
+ break;
+ }
+ }
+ else if (gfc_match ("ancestor : ") == MATCH_YES)
+ {
+ gfc_error ("sorry, unimplemented: 'ancestor' not yet "
+ "supported at %C");
+ break;
+ }
+ else if (gfc_match ("%e )", &c->device) == MATCH_YES)
+ {
+ }
+ else
+ {
+ gfc_error ("Expected integer expression or a single device-"
+ "modifier %<device_num%> or %<ancestor%> at %C");
+ break;
+ }
+ continue;
+ }
if ((mask & OMP_CLAUSE_DEVICE)
&& openacc
&& gfc_match ("device ( ") == MATCH_YES
diff --git a/gcc/testsuite/c-c++-common/gomp/target-device-1.c
b/gcc/testsuite/c-c++-common/gomp/target-device-1.c
new file mode 100644
index 0000000..6b01e4b
--- /dev/null
+++ b/gcc/testsuite/c-c++-common/gomp/target-device-1.c
@@ -0,0 +1,78 @@
+/* { dg-do compile } */
+
+void
+foo (void)
+{
+ /* Test to ensure that the device modifiers are parsed correctly in device
clauses. */
+
+ int n;
+
+ #pragma omp target device (1)
+ ;
+
+ #pragma omp target device (n)
+ ;
+
+ #pragma omp target device (n + 1)
+ ;
+
+ #pragma omp target device (device_num : 1)
+ ;
+
+ #pragma omp target device (device_num : n)
+ ;
+
+ #pragma omp target device (device_num : n + 1)
+ ;
+
+ #pragma omp target device (ancestor : 1) /* { dg-message "sorry,
unimplemented: 'ancestor' not yet supported" } */
+ ;
+
+ #pragma omp target device (ancestor : n) /* { dg-message "sorry,
unimplemented: 'ancestor' not yet supported" } */
+ ;
+
+ #pragma omp target device (ancestor : n + 1) /* { dg-message "sorry,
unimplemented: 'ancestor' not yet supported" } */
+ ;
+
+ #pragma omp target device (invalid : 1) /* { dg-error "expected
device-modifier 'ancestor' or 'device_num'" } */
+ ;
+
+ #pragma omp target device ( : 1) /* { dg-error "expected device-modifier
'ancestor' or 'device_num'" } */
+ ;
+
+ #pragma omp target device ( , : 1) /* { dg-error "expected device-modifier
'ancestor' or 'device_num'" } */
+ ;
+
+ #pragma omp target device (ancestor, device_num : 1) /* { dg-error "expected
only one device-modifier 'ancestor' or 'device_num'" } */
+ ;
+
+ #pragma omp target device (ancestor, device_num, ancestor : 1) /* { dg-error
"expected only one device-modifier 'ancestor' or 'device_num'" } */
+ ;
+
+ #pragma omp target device (device_num device_num : 1) /* { dg-error
"expected only one device-modifier 'ancestor' or 'device_num'" } */
+ ;
+
+ #pragma omp target device (ancestor device_num : 1) /* { dg-error "expected
only one device-modifier 'ancestor' or 'device_num'" } */
+ ;
+
+ #pragma omp target device (device_num, invalid : 1) /* { dg-error "expected
only one device-modifier 'ancestor' or 'device_num'" } */
+ ;
+
+ #pragma omp target device (ancestor, invalid : 1) /* { dg-error "expected
only one device-modifier 'ancestor' or 'device_num'" } */
+ ;
+
+ #pragma omp target device (ancestor, , , : 1) /* { dg-error "expected only
one device-modifier 'ancestor' or 'device_num'" } */
+ ;
+
+ #pragma omp target device (invalid, ancestor : 1) /* { dg-error "expected
device-modifier 'ancestor' or 'device_num'" } */
+ ;
+
+ #pragma omp target device (invalid, invalid, ancestor : 1) /* { dg-error
"expected device-modifier 'ancestor' or 'device_num'" } */
+ ;
+
+ #pragma omp target device (device_num invalid : 1) /* { dg-error "expected
only one device-modifier 'ancestor' or 'device_num'" } */
+ ;
+
+ #pragma omp target device (device_num : n, n) /* { dg-error "expected '\\)'
before ','" } */
+ ;
+}
diff --git a/gcc/testsuite/c-c++-common/gomp/target-device-2.c
b/gcc/testsuite/c-c++-common/gomp/target-device-2.c
new file mode 100644
index 0000000..69c84d0
--- /dev/null
+++ b/gcc/testsuite/c-c++-common/gomp/target-device-2.c
@@ -0,0 +1,13 @@
+/* { dg-do compile } */
+/* { dg-additional-options "-fdump-tree-original" } */
+
+void
+foo (void)
+{
+ /* Test to ensure that device-modifier 'device_num' is parsed correctly in
device clauses. */
+
+ #pragma omp target device (device_num : 42)
+ ;
+}
+
+/* { dg-final { scan-tree-dump "pragma omp target \[^\n\r)]*device\\(42\\)"
"original" } } */
diff --git a/gcc/testsuite/gfortran.dg/gomp/target-device-1.f90
b/gcc/testsuite/gfortran.dg/gomp/target-device-1.f90
new file mode 100644
index 0000000..d4e31e9
--- /dev/null
+++ b/gcc/testsuite/gfortran.dg/gomp/target-device-1.f90
@@ -0,0 +1,76 @@
+! { dg-do compile }
+
+implicit none
+
+integer :: n
+
+!$omp target device (1)
+!$omp end target
+
+!$omp target device (n)
+!$omp end target
+
+!$omp target device (n + 1)
+!$omp end target
+
+!$omp target device (device_num : 1)
+!$omp end target
+
+!$omp target device (device_num : n)
+!$omp end target
+
+!$omp target device (device_num : n + 1)
+!$omp end target
+
+!$omp target device (ancestor : 1) ! { dg-error "sorry, unimplemented:
'ancestor' not yet supported" }
+! !$omp end target
+
+!$omp target device (ancestor : n) ! { dg-error "sorry, unimplemented:
'ancestor' not yet supported" }
+! !$omp end target
+
+!$omp target device (ancestor : n + 1) ! { dg-error "sorry, unimplemented:
'ancestor' not yet supported" }
+! !$omp end target
+
+!$omp target device (invalid : 1) ! { dg-error "Expected integer expression
or a single device-modifier 'device_num' or 'ancestor' at" }
+! !$omp end target
+
+!$omp target device ( : 1) ! { dg-error "Expected integer expression or a
single device-modifier 'device_num' or 'ancestor' at" }
+! !$omp end target
+
+!$omp target device ( , : 1) ! { dg-error "Expected integer expression or a
single device-modifier 'device_num' or 'ancestor' at" }
+! !$omp end target
+
+!$omp target device (ancestor, device_num : 1) ! { dg-error "Expected integer
expression or a single device-modifier 'device_num' or 'ancestor' at" }
+! !$omp end target
+
+!$omp target device (ancestor, device_num, ancestor : 1) ! { dg-error
"Expected integer expression or a single device-modifier 'device_num' or
'ancestor' at" }
+! !$omp end target
+
+!$omp target device (device_num device_num : 1) ! { dg-error "Expected
integer expression or a single device-modifier 'device_num' or 'ancestor' at" }
+! !$omp end target
+
+!$omp target device (ancestor device_num : 1) ! { dg-error "Expected integer
expression or a single device-modifier 'device_num' or 'ancestor' at" }
+! !$omp end target
+
+!$omp target device (device_num, invalid : 1) ! { dg-error "Expected integer
expression or a single device-modifier 'device_num' or 'ancestor' at" }
+! !$omp end target
+
+!$omp target device (ancestor, invalid : 1) ! { dg-error "Expected integer
expression or a single device-modifier 'device_num' or 'ancestor' at" }
+! !$omp end target
+
+!$omp target device (ancestor, , , : 1) ! { dg-error "Expected integer
expression or a single device-modifier 'device_num' or 'ancestor' at" }
+! !$omp end target
+
+!$omp target device (invalid, ancestor : 1) ! { dg-error "xpected integer
expression or a single device-modifier 'device_num' or 'ancestor' at" }
+! !$omp end target
+
+!$omp target device (invalid, invalid, ancestor : 1) ! { dg-error "xpected
integer expression or a single device-modifier 'device_num' or 'ancestor' at" }
+! !$omp end target
+
+!$omp target device (device_num invalid : 1) ! { dg-error "Expected integer
expression or a single device-modifier 'device_num' or 'ancestor' at" }
+! !$omp end target
+
+!$omp target device (device_num : n, n) ! { dg-error "Expected integer
expression at" }
+! !$omp end target
+
+end
diff --git a/gcc/testsuite/gfortran.dg/gomp/target-device-2.f90
b/gcc/testsuite/gfortran.dg/gomp/target-device-2.f90
new file mode 100644
index 0000000..40fa2d8
--- /dev/null
+++ b/gcc/testsuite/gfortran.dg/gomp/target-device-2.f90
@@ -0,0 +1,13 @@
+! { dg-do compile }
+! { dg-additional-options "-fdump-tree-original" }
+
+implicit none
+
+integer :: n
+
+!$omp target device (device_num : 42)
+!$omp end target
+
+end
+
+! { dg-final { scan-tree-dump "pragma omp target \[^\n\r)]*device\\(42\\)"
"original" } }
