diff --git a/gcc/c/c-parser.c b/gcc/c/c-parser.c
index 356cf2504d4b4a56bd5de176a9845f07fc3b551c..3b1d10f1a36760a64a7d580b79dffc733c1c5eef 100644
--- a/gcc/c/c-parser.c
+++ b/gcc/c/c-parser.c
@@ -15945,37 +15945,87 @@ 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;
- matching_parens parens;
- if (parens.require_open (parser))
- {
- 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);
+ location_t expr_loc;
+ c_expr expr;
+ tree c, t;
+ bool ancestor = false;
- parens.skip_until_found_close (parser);
+ matching_parens parens;
+ if (!parens.require_open (parser))
+ return list;
- if (!INTEGRAL_TYPE_P (TREE_TYPE (t)))
+ if (c_parser_next_token_is (parser, CPP_NAME)
+ && c_parser_peek_2nd_token (parser)->type == CPP_COLON)
+ {
+ c_token *tok = c_parser_peek_token (parser);
+ const char *p = IDENTIFIER_POINTER (tok->value);
+ if (strcmp ("ancestor", p) == 0)
{
- c_parser_error (parser, "expected integer expression");
+ /* A requires directive with the reverse_offload clause must be
+ specified. */
+ if ((omp_requires_mask & OMP_REQUIRES_REVERSE_OFFLOAD) == 0)
+ {
+ error_at (tok->location, "%<ancestor%> device modifier not "
+ "preceded by %<requires%> directive "
+ "with %<reverse_offload%> clause");
+ parens.skip_until_found_close (parser);
+ return list;
+ }
+ ancestor = true;
+ }
+ else if (strcmp ("device_num", p) == 0)
+ ;
+ else
+ {
+ error_at (tok->location, "expected %<ancestor%> or %<device_num%>");
+ parens.skip_until_found_close (parser);
return list;
}
+ c_parser_consume_token (parser);
+ c_parser_consume_token (parser);
+ }
- 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);
+ 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;
}
+ if (ancestor && TREE_CODE (t) == INTEGER_CST && !integer_onep (t))
+ {
+ error_at (expr_loc, "the %<device%> clause expression must evaluate to "
+ "%<1%>");
+ 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;
+ OMP_CLAUSE_DEVICE_ANCESTOR (c) = ancestor;
+
+ list = c;
return list;
}
diff --git a/gcc/cp/parser.c b/gcc/cp/parser.c
index 797e70ba5bbfd92ef1e1451ee4be4ef135a91943..7dc4eae7102537fb614efba58b2d8d58a05f12cf 100644
--- a/gcc/cp/parser.c
+++ b/gcc/cp/parser.c
@@ -39049,18 +39049,57 @@ 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,
location_t location)
{
tree t, c;
+ bool ancestor = false;
matching_parens parens;
if (!parens.require_open (parser))
return list;
+ if (cp_lexer_next_token_is (parser->lexer, CPP_NAME)
+ && cp_lexer_nth_token_is (parser->lexer, 2, CPP_COLON))
+ {
+ cp_token *tok = cp_lexer_peek_token (parser->lexer);
+ const char *p = IDENTIFIER_POINTER (tok->u.value);
+ if (strcmp ("ancestor", p) == 0)
+ {
+ ancestor = true;
+
+ /* A requires directive with the reverse_offload clause must be
+ specified. */
+ if ((omp_requires_mask & OMP_REQUIRES_REVERSE_OFFLOAD) == 0)
+ {
+ error_at (tok->location, "%<ancestor%> device modifier not "
+ "preceded by %<requires%> directive "
+ "with %<reverse_offload%> clause");
+ cp_parser_skip_to_closing_parenthesis (parser, true, false, true);
+ return list;
+ }
+ }
+ else if (strcmp ("device_num", p) == 0)
+ ;
+ else
+ {
+ error_at (tok->location, "expected %<ancestor%> or %<device_num%>");
+ cp_parser_skip_to_closing_parenthesis (parser, true, false, true);
+ return list;
+ }
+ cp_lexer_consume_token (parser->lexer);
+ cp_lexer_consume_token (parser->lexer);
+ }
+
t = cp_parser_assignment_expression (parser);
if (t == error_mark_node
@@ -39075,6 +39114,7 @@ cp_parser_omp_clause_device (cp_parser *parser, tree list,
c = build_omp_clause (location, OMP_CLAUSE_DEVICE);
OMP_CLAUSE_DEVICE_ID (c) = t;
OMP_CLAUSE_CHAIN (c) = list;
+ OMP_CLAUSE_DEVICE_ANCESTOR (c) = ancestor;
return c;
}
diff --git a/gcc/cp/semantics.c b/gcc/cp/semantics.c
index e191aa36c98fa4ea88bfb88bd3f94fa31e2c0bc8..f4b042fb676bbd088b455e6d8774b05088170450 100644
--- a/gcc/cp/semantics.c
+++ b/gcc/cp/semantics.c
@@ -7334,6 +7334,15 @@ finish_omp_clauses (tree clauses, enum c_omp_region_type ort)
"%<device%> id must be integral");
remove = true;
}
+ else if (OMP_CLAUSE_DEVICE_ANCESTOR (c)
+ && TREE_CODE (t) == INTEGER_CST
+ && !integer_onep (t))
+ {
+ error_at (OMP_CLAUSE_LOCATION (c),
+ "the %<device%> clause expression must evaluate to "
+ "%<1%>");
+ remove = true;
+ }
else
{
t = mark_rvalue_use (t);
diff --git a/gcc/fortran/gfortran.h b/gcc/fortran/gfortran.h
index 48cdcdf6cb87c539602d1f5786bb593108f2bbaa..fdf556eef3d1a24411d7a750ab522a1c2ec2cf78 100644
--- a/gcc/fortran/gfortran.h
+++ b/gcc/fortran/gfortran.h
@@ -1482,6 +1482,7 @@ typedef struct gfc_omp_clauses
struct gfc_expr *dist_chunk_size;
struct gfc_expr *message;
const char *critical_name;
+ bool ancestor;
enum gfc_omp_default_sharing default_sharing;
enum gfc_omp_atomic_op atomic_op;
enum gfc_omp_defaultmap defaultmap[OMP_DEFAULTMAP_CAT_NUM];
diff --git a/gcc/fortran/openmp.c b/gcc/fortran/openmp.c
index 715fd32151202c893915e40dc616e1165c852d8b..64ecd547d267ffb8949c0d2c5063791db9c22931 100644
--- a/gcc/fortran/openmp.c
+++ b/gcc/fortran/openmp.c
@@ -1825,11 +1825,54 @@ gfc_match_omp_clauses (gfc_omp_clauses **cp, const omp_mask mask,
continue;
if ((mask & OMP_CLAUSE_DEVICE)
&& !openacc
- && (m = gfc_match_dupl_check (!c->device, "device", true,
- &c->device)) != MATCH_NO)
+ && ((m = gfc_match_dupl_check (!c->device, "device", true))
+ != MATCH_NO))
{
if (m == MATCH_ERROR)
goto error;
+ c->ancestor = false;
+ 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)
+ {
+ c->ancestor = true;
+ if (!(gfc_current_ns->omp_requires & OMP_REQ_REVERSE_OFFLOAD))
+ {
+ gfc_error ("%<ancestor%> device modifier not "
+ "preceded by %<requires%> directive "
+ "with %<reverse_offload%> clause at %C");
+ break;
+ }
+ locus old_loc2 = gfc_current_locus;
+ if (gfc_match ("%e )", &c->device) == MATCH_YES)
+ {
+ int device = 0;
+ if (!gfc_extract_int (c->device, &device) && device != 1)
+ {
+ gfc_current_locus = old_loc2;
+ gfc_error ("the %<device%> clause expression must "
+ "evaluate to %<1%> at %C");
+ break;
+ }
+ }
+ else
+ {
+ gfc_error ("Expected integer expression at %C");
+ break;
+ }
+ }
+ else if (gfc_match ("%e )", &c->device) != MATCH_YES)
+ {
+ gfc_error ("Expected integer expression or a single device-"
+ "modifier %<device_num%> or %<ancestor%> at %C");
+ break;
+ }
continue;
}
if ((mask & OMP_CLAUSE_DEVICE)
diff --git a/gcc/fortran/trans-openmp.c b/gcc/fortran/trans-openmp.c
index 40d2fd206e43afdde2b06f6ee331f10e879874c7..6f9b0e390dea0eaec3f8c27bcee7f8dc8b26081e 100644
--- a/gcc/fortran/trans-openmp.c
+++ b/gcc/fortran/trans-openmp.c
@@ -3950,6 +3950,10 @@ gfc_trans_omp_clauses (stmtblock_t *block, gfc_omp_clauses *clauses,
c = build_omp_clause (gfc_get_location (&where), OMP_CLAUSE_DEVICE);
OMP_CLAUSE_DEVICE_ID (c) = device;
+
+ if (clauses->ancestor)
+ OMP_CLAUSE_DEVICE_ANCESTOR (c) = 1;
+
omp_clauses = gfc_trans_add_clause (c, omp_clauses);
}
diff --git a/gcc/gimplify.c b/gcc/gimplify.c
index 070d0e4df45a6f8ba12c5edbf7790443d10931cb..400b6f0f4c186bae6df4bb250596375a12e02e0a 100644
--- a/gcc/gimplify.c
+++ b/gcc/gimplify.c
@@ -10107,6 +10107,38 @@ gimplify_scan_omp_clauses (tree *list_p, gimple_seq *pre_p,
case OMP_CLAUSE_THREAD_LIMIT:
case OMP_CLAUSE_DIST_SCHEDULE:
case OMP_CLAUSE_DEVICE:
+ if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_DEVICE
+ && OMP_CLAUSE_DEVICE_ANCESTOR (c))
+ {
+ if (code != OMP_TARGET)
+ {
+ error_at (OMP_CLAUSE_LOCATION (c),
+ "%<device%> clause with %<ancestor%> is only "
+ "allowed on %<target%> construct");
+ remove = true;
+ break;
+ }
+
+ tree clauses = *orig_list_p;
+ for (; clauses ; clauses = OMP_CLAUSE_CHAIN (clauses))
+ if (OMP_CLAUSE_CODE (clauses) != OMP_CLAUSE_DEVICE
+ && OMP_CLAUSE_CODE (clauses) != OMP_CLAUSE_FIRSTPRIVATE
+ && OMP_CLAUSE_CODE (clauses) != OMP_CLAUSE_PRIVATE
+ && OMP_CLAUSE_CODE (clauses) != OMP_CLAUSE_DEFAULTMAP
+ && OMP_CLAUSE_CODE (clauses) != OMP_CLAUSE_MAP
+ )
+ {
+ error_at (OMP_CLAUSE_LOCATION (c),
+ "with %<ancestor%>, only the %<device%>, "
+ "%<firstprivate%>, %<private%>, %<defaultmap%>, "
+ "and %<map%> clauses may appear on the "
+ "construct");
+ remove = true;
+ break;
+ }
+ }
+ /* Fall through. */
+
case OMP_CLAUSE_PRIORITY:
case OMP_CLAUSE_GRAINSIZE:
case OMP_CLAUSE_NUM_TASKS:
diff --git a/gcc/omp-expand.c b/gcc/omp-expand.c
index 66c64f5a37b75e0e63a8e70d47621150170a2826..ecb8ee517805a25d488e29a2ffecc972bab481c9 100644
--- a/gcc/omp-expand.c
+++ b/gcc/omp-expand.c
@@ -9613,6 +9613,8 @@ expand_omp_target (struct omp_region *region)
{
device = OMP_CLAUSE_DEVICE_ID (c);
device_loc = OMP_CLAUSE_LOCATION (c);
+ if (OMP_CLAUSE_DEVICE_ANCESTOR (c))
+ sorry_at (device_loc, "%<ancestor%> not yet supported");
}
else
{
diff --git a/gcc/omp-low.c b/gcc/omp-low.c
index a0b41afa3ebff5e04b0d515b4241a2ab5e67a4d9..6686946e5a7b688e7aedfa61d953b1caf3f0b154 100644
--- a/gcc/omp-low.c
+++ b/gcc/omp-low.c
@@ -3114,6 +3114,16 @@ check_omp_nesting_restrictions (gimple *stmt, omp_context *ctx)
if (gimple_code (ctx->stmt) == GIMPLE_OMP_TARGET
&& gimple_omp_target_kind (ctx->stmt) == GF_OMP_TARGET_KIND_REGION)
{
+ c = omp_find_clause (gimple_omp_target_clauses (ctx->stmt),
+ OMP_CLAUSE_DEVICE);
+ if (c && OMP_CLAUSE_DEVICE_ANCESTOR (c))
+ {
+ error_at (gimple_location (stmt),
+ "OpenMP constructs are not allowed in target region "
+ "with %<ancestor%>");
+ return false;
+ }
+
if (gimple_code (stmt) == GIMPLE_OMP_TEAMS && !ctx->teams_nested_p)
ctx->teams_nested_p = true;
else
@@ -4063,6 +4073,17 @@ scan_omp_1_stmt (gimple_stmt_iterator *gsi, bool *handled_ops_p,
"OpenMP runtime API call %qD in a region with "
"%<order(concurrent)%> clause", fndecl);
}
+ if (gimple_code (ctx->stmt) == GIMPLE_OMP_TARGET
+ && (gimple_omp_target_kind (ctx->stmt)
+ == GF_OMP_TARGET_KIND_REGION))
+ {
+ tree tgt_clauses = gimple_omp_target_clauses (ctx->stmt);
+ tree c = omp_find_clause (tgt_clauses, OMP_CLAUSE_DEVICE);
+ if (c && OMP_CLAUSE_DEVICE_ANCESTOR (c))
+ error_at (gimple_location (stmt),
+ "OpenMP runtime API call %qD in a region with "
+ "%<device(ancestor)%> clause", fndecl);
+ }
}
}
}
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 0000000000000000000000000000000000000000..98228626df9ca7e48d5081bf49398d5a28571ca7
--- /dev/null
+++ b/gcc/testsuite/c-c++-common/gomp/target-device-1.c
@@ -0,0 +1,32 @@
+/* { dg-do compile } */
+
+void
+foo (int n)
+{
+ /* Test to ensure that 'device_num' is parsed correctly in device clauses. */
+
+ #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 (invalid : 1) /* { dg-error "expected 'ancestor' or 'device_num'" "" { target *-*-* } } */
+ /* { dg-error "expected '\\)' before 'invalid'" "" { target c } .-1 } */
+ ;
+
+ #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 0000000000000000000000000000000000000000..b711ea15a24654a4d7c621bb9ec7c9ccd7dd3cfb
--- /dev/null
+++ b/gcc/testsuite/c-c++-common/gomp/target-device-2.c
@@ -0,0 +1,14 @@
+/* { dg-do compile } */
+/* { dg-additional-options "-fdump-tree-original" } */
+
+ /* Test to ensure that device-modifier 'device_num' is parsed correctly in
+ device clauses. */
+
+void
+foo (void)
+{
+ #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/c-c++-common/gomp/target-device-ancestor-1.c b/gcc/testsuite/c-c++-common/gomp/target-device-ancestor-1.c
new file mode 100644
index 0000000000000000000000000000000000000000..b3c1ce89430606bfd8d3db74e43280cf82d86e95
--- /dev/null
+++ b/gcc/testsuite/c-c++-common/gomp/target-device-ancestor-1.c
@@ -0,0 +1,13 @@
+/* { dg-do compile } */
+
+void
+foo (void)
+{
+ /* Ensure that a 'requires' directive with the 'reverse_offload' clause was
+ specified. */
+
+ #pragma omp target device (ancestor : 1) /* { dg-error "'ancestor' device modifier not preceded by 'requires' directive with 'reverse_offload' clause" } */
+ /* { dg-error "expected '\\)' before 'ancestor'" "" { target c } .-1 } */
+
+ ;
+}
diff --git a/gcc/testsuite/c-c++-common/gomp/target-device-ancestor-2.c b/gcc/testsuite/c-c++-common/gomp/target-device-ancestor-2.c
new file mode 100644
index 0000000000000000000000000000000000000000..cf05c505004b3352d4562c6471376859c6957c34
--- /dev/null
+++ b/gcc/testsuite/c-c++-common/gomp/target-device-ancestor-2.c
@@ -0,0 +1,82 @@
+/* { dg-do compile } */
+
+#pragma omp requires reverse_offload /* { dg-message "sorry, unimplemented: 'reverse_offload' clause on 'requires' directive not supported yet" } */
+
+void
+foo (int n)
+{
+ /* The following test is marked with 'xfail' because a previous 'sorry' from
+ 'reverse_offload' suppresses the 'sorry' for 'ancestor'. */
+ #pragma omp target device (ancestor: 1) /* { dg-message "" "sorry, unimplemented: 'ancestor' not yet supported" { xfail *-*-* } } */
+ ;
+
+
+ /* Ensure that the integer expression in the 'device' clause for
+ device-modifier 'ancestor' evaluates to '1' in case of a constant. */
+
+ #pragma omp target device (ancestor : 1)
+ ;
+ #pragma omp target device (ancestor : 42) /* { dg-error "the 'device' clause expression must evaluate to '1'" } */
+ ;
+
+ #pragma omp target device (ancestor : n) /* { dg-message "" "sorry, unimplemented: 'ancestor' not yet supported" { xfail *-*-* } } */
+ ;
+ #pragma omp target device (ancestor : n + 1) /* { dg-message "" "sorry, unimplemented: 'ancestor' not yet supported" { xfail *-*-* } } */
+ ;
+
+
+ /* Ensure that only one 'device' clause appears on the construct. */
+
+ #pragma omp target device (17) device (42) /* { dg-error "too many 'device' clauses" } */
+ ;
+
+
+ /* Ensure that with 'ancestor' only the 'device', 'firstprivate', 'private',
+ 'defaultmap', and 'map' clauses appear on the construct. */
+
+ #pragma omp target nowait device (ancestor: 1) /* { dg-error "with 'ancestor', only the 'device', 'firstprivate', 'private', 'defaultmap', and 'map' clauses may appear on the construct" } */
+ ;
+ #pragma omp target device (ancestor: 1) nowait /* { dg-error "with 'ancestor', only the 'device', 'firstprivate', 'private', 'defaultmap', and 'map' clauses may appear on the construct" } */
+ ;
+ #pragma omp target nowait device (42)
+ ;
+ #pragma omp target nowait device (device_num: 42)
+ ;
+
+ int a = 0, b = 0, c = 0;
+ #pragma omp target device (ancestor: 1) firstprivate (a) private (b) defaultmap (none) map (c)
+ ;
+
+
+ /* Ensure that 'ancestor' is only used with 'target' constructs (not with
+ 'target data', 'target update' etc.). */
+
+ #pragma omp target data map (a) device (ancestor: 1) /* { dg-error "'device' clause with 'ancestor' is only allowed on 'target' construct" } */
+ ;
+ #pragma omp target enter data map (to: a) device (ancestor: 1) /* { dg-error "'device' clause with 'ancestor' is only allowed on 'target' construct" } */
+ #pragma omp target exit data map (from: a) device (ancestor: 1) /* { dg-error "'device' clause with 'ancestor' is only allowed on 'target' construct" } */
+ #pragma omp target update to (a) device (ancestor: 1) /* { dg-error "'device' clause with 'ancestor' is only allowed on 'target' construct" "" { target *-*-* } } */
+
+
+ /* Ensure that no OpenMP constructs appear inside target regions with
+ 'ancestor'. */
+
+ #pragma omp target device (ancestor: 1)
+ {
+ #pragma omp teams /* { dg-error "OpenMP constructs are not allowed in target region with 'ancestor'" } */
+ ;
+ }
+
+ #pragma omp target device (device_num: 1)
+ {
+ #pragma omp teams
+ ;
+ }
+
+ #pragma omp target device (1)
+ {
+ #pragma omp teams
+ ;
+ }
+
+}
diff --git a/gcc/testsuite/c-c++-common/gomp/target-device-ancestor-3.c b/gcc/testsuite/c-c++-common/gomp/target-device-ancestor-3.c
new file mode 100644
index 0000000000000000000000000000000000000000..5e3a478fd5b4d42bf63f71eedea32c4b546124e5
--- /dev/null
+++ b/gcc/testsuite/c-c++-common/gomp/target-device-ancestor-3.c
@@ -0,0 +1,37 @@
+#ifdef __cplusplus
+extern "C" {
+#endif
+
+int omp_get_num_teams (void);
+
+#ifdef __cplusplus
+}
+#endif
+
+/* { dg-do compile } */
+
+#pragma omp requires reverse_offload /* { dg-message "sorry, unimplemented: 'reverse_offload' clause on 'requires' directive not supported yet" } */
+
+void
+foo (void)
+{
+ /* Ensure that no calls to OpenMP API runtime routines are allowed inside the
+ corresponding target region. */
+
+ int a;
+
+ #pragma omp target device (ancestor: 1)
+ {
+ a = omp_get_num_teams (); /* { dg-error "OpenMP runtime API call '\[^\n\r]*omp_get_num_teams\[^\n\r]*' in a region with 'device\\(ancestor\\)' clause" } */
+ }
+
+ #pragma omp target device (device_num: 1)
+ {
+ a = omp_get_num_teams ();
+ }
+
+ #pragma omp target device (1)
+ {
+ a = omp_get_num_teams ();
+ }
+}
diff --git a/gcc/testsuite/c-c++-common/gomp/target-device-ancestor-4.c b/gcc/testsuite/c-c++-common/gomp/target-device-ancestor-4.c
new file mode 100644
index 0000000000000000000000000000000000000000..b4b5620bbc0f90ab93b33a1b9b10911492b2d541
--- /dev/null
+++ b/gcc/testsuite/c-c++-common/gomp/target-device-ancestor-4.c
@@ -0,0 +1,17 @@
+/* { dg-do compile } */
+/* { dg-additional-options "-fdump-tree-original" } */
+
+ /* Test to ensure that device-modifier 'ancestor' is parsed correctly in
+ device clauses. */
+
+#pragma omp requires reverse_offload /* { dg-message "sorry, unimplemented: 'reverse_offload' clause on 'requires' directive not supported yet" } */
+
+void
+foo (void)
+{
+ #pragma omp target device (ancestor: 1) /* { dg-message "" "sorry, unimplemented: 'ancestor' not yet supported" { xfail *-*-* } } */
+ ;
+
+}
+
+/* { dg-final { scan-tree-dump "pragma omp target \[^\n\r)]*device\\(ancestor:1\\)" "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 0000000000000000000000000000000000000000..20b97559978e6a783f5e3bf8f8146e3a355cb92e
--- /dev/null
+++ b/gcc/testsuite/gfortran.dg/gomp/target-device-1.f90
@@ -0,0 +1,67 @@
+! { 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 (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" }
+! !$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 0000000000000000000000000000000000000000..133b805b8e1a07fb37212b5d4282d5f87e622ca0
--- /dev/null
+++ b/gcc/testsuite/gfortran.dg/gomp/target-device-2.f90
@@ -0,0 +1,12 @@
+! { dg-do compile }
+! { dg-additional-options "-fdump-tree-original" }
+
+! Test to ensure that device-modifier 'device_num' is parsed correctly in
+! device clauses.
+
+!$omp target device (device_num : 42)
+!$omp end target
+
+end
+
+! { dg-final { scan-tree-dump "pragma omp target \[^\n\r)]*device\\(42\\)" "original" } }
diff --git a/gcc/testsuite/gfortran.dg/gomp/target-device-ancestor-1.f90 b/gcc/testsuite/gfortran.dg/gomp/target-device-ancestor-1.f90
new file mode 100644
index 0000000000000000000000000000000000000000..9a170dbee07dcda63313134ebc0901c22301af3d
--- /dev/null
+++ b/gcc/testsuite/gfortran.dg/gomp/target-device-ancestor-1.f90
@@ -0,0 +1,9 @@
+! { dg-do compile }
+
+! Ensure that a 'requires' directive with the 'reverse_offload' clause was
+! specified.
+
+!$omp target device (ancestor:1) ! { dg-error "'ancestor' device modifier not preceded by 'requires' directive with 'reverse_offload' clause" }
+! !$omp end target
+
+end
\ No newline at end of file
diff --git a/gcc/testsuite/gfortran.dg/gomp/target-device-ancestor-2.f90 b/gcc/testsuite/gfortran.dg/gomp/target-device-ancestor-2.f90
new file mode 100644
index 0000000000000000000000000000000000000000..117a1d000a5748fb6784168132ea6f9458bd1d4d
--- /dev/null
+++ b/gcc/testsuite/gfortran.dg/gomp/target-device-ancestor-2.f90
@@ -0,0 +1,92 @@
+! { dg-do compile }
+
+implicit none
+
+integer :: a, b, c
+
+!$omp requires reverse_offload ! { dg-error "Sorry, 'reverse_offload' clause at \\(1\\) on REQUIRES directive is not yet supported" }
+
+
+! The following test case is marked with 'xfail' because a previous 'sorry' from
+! 'reverse_offload' suppresses the 'sorry' for 'ancestor'.
+
+!$omp target device (ancestor: 1) ! { dg-message "" "sorry, unimplemented: 'ancestor' not yet supported" { xfail *-*-* } }
+!$omp end target
+
+!$omp target device (ancestor : a) ! { dg-message "" "sorry, unimplemented: 'ancestor' not yet supported" { xfail *-*-* } }
+!$omp end target
+
+!$omp target device (ancestor : a + 1) ! { dg-message "" "sorry, unimplemented: 'ancestor' not yet supported" { xfail *-*-* } }
+!$omp end target
+
+
+! Ensure that the integer expression in the 'device' clause for
+! device-modifier 'ancestor' evaluates to '1' in case of a constant.
+
+!$omp target device (ancestor: 42) ! { dg-error "the 'device' clause expression must evaluate to '1'" }
+! !$omp end target
+
+!$omp target device (device_num:42)
+!$omp end target
+
+!$omp target device (42)
+!$omp end target
+
+
+! Ensure that no OpenMP constructs appear inside target regions with 'ancestor'.
+! The following test case is marked with 'xfail' because a previous 'sorry' from
+! 'reverse_offload' suppresses the 'sorry' for 'ancestor'.
+
+!$omp target device (ancestor: 1)
+ !$omp teams ! { dg-error "" "OpenMP constructs are not allowed in target region with 'ancestor'" { xfail *-*-* } }
+ !$omp end teams
+!$omp end target
+
+!$omp target device (device_num: 1)
+ !$omp teams
+ !$omp end teams
+!$omp end target
+
+!$omp target device (1)
+ !$omp teams
+ !$omp end teams
+!$omp end target
+
+
+! Ensure that with 'ancestor' only the 'device', 'firstprivate', 'private',
+! 'defaultmap', and 'map' clauses appear on the construct.
+! The following test case is marked with 'xfail' because a previous 'sorry' from
+! 'reverse_offload' suppresses the 'sorry' for 'ancestor'.
+
+!$omp target nowait device (ancestor: 1) ! { dg-error "" "with 'ancestor', only the 'device', 'firstprivate', 'private', 'defaultmap', and 'map' clauses may appear on the construct" { xfail *-*-* } }
+!$omp end target
+
+!$omp target device (ancestor: 1) nowait ! { dg-error "" "with 'ancestor', only the 'device', 'firstprivate', 'private', 'defaultmap', and 'map' clauses may appear on the construct" { xfail *-*-* } }
+!$omp end target
+
+!$omp target nowait device (device_num: 1)
+!$omp end target
+
+!$omp target nowait device (1)
+!$omp end target
+
+!$omp target device (ancestor: 1) firstprivate (a) private (b) defaultmap (none) map (c)
+!$omp end target
+
+
+! Ensure that 'ancestor' is only used with 'target' constructs (not with
+! 'target data', 'target update' etc.).
+! The following test case is marked with 'xfail' because a previous 'sorry' from
+! 'reverse_offload' suppresses the 'sorry' for 'ancestor'.
+
+!$omp target data map (a) device (ancestor: 1) ! { dg-error "" "'device' clause with 'ancestor' is only allowed on 'target' construct" { xfail *-*-* } }
+!$omp end target data
+
+!$omp target enter data map (to: a) device (ancestor: 1) ! { dg-error "" "'device' clause with 'ancestor' is only allowed on 'target' construct" { xfail *-*-* } }
+!$omp target exit data map (from: a) device (ancestor: 1) ! { dg-error "" "'device' clause with 'ancestor' is only allowed on 'target' construct" { xfail *-*-* } }
+
+!$omp target update to (a) device (ancestor: 1) ! { dg-error "'device' clause with 'ancestor' is only allowed on 'target' construct" "" { xfail *-*-* } }
+! { dg-error "with 'ancestor', only the 'device', 'firstprivate', 'private', 'defaultmap', and 'map' clauses may appear on the construct" "" { xfail *-*-* } .-1 }
+
+
+end
\ No newline at end of file
diff --git a/gcc/testsuite/gfortran.dg/gomp/target-device-ancestor-3.f90 b/gcc/testsuite/gfortran.dg/gomp/target-device-ancestor-3.f90
new file mode 100644
index 0000000000000000000000000000000000000000..f1145bde2ece0bee085c97f68de4b64be032b984
--- /dev/null
+++ b/gcc/testsuite/gfortran.dg/gomp/target-device-ancestor-3.f90
@@ -0,0 +1,33 @@
+! { dg-do compile }
+
+! This testcase ensure that no calls to OpenMP API runtime routines are allowed
+! inside the corresponding target region.
+
+module my_omp_mod
+ use iso_c_binding
+ interface
+ integer function omp_get_thread_num ()
+ end
+ end interface
+end
+
+subroutine f1 ()
+ use my_omp_mod
+ implicit none
+ integer :: n
+
+ !$omp requires reverse_offload ! { dg-error "Sorry, 'reverse_offload' clause at \\(1\\) on REQUIRES directive is not yet supported" }
+
+ !$omp target device (ancestor : 1)
+ n = omp_get_thread_num () ! { dg-error "" "OpenMP runtime API call 'omp_get_thread_num' in a region with 'device\\(ancestor\\)' clause" { xfail *-*-* } }
+ !$omp end target
+
+ !$omp target device (device_num : 1)
+ n = omp_get_thread_num ()
+ !$omp end target
+
+ !$omp target device (1)
+ n = omp_get_thread_num ()
+ !$omp end target
+
+end
\ No newline at end of file
diff --git a/gcc/testsuite/gfortran.dg/gomp/target-device-ancestor-4.f90 b/gcc/testsuite/gfortran.dg/gomp/target-device-ancestor-4.f90
new file mode 100644
index 0000000000000000000000000000000000000000..540b3d0355c08d454fd7ffb82df9fe6129b3e1f1
--- /dev/null
+++ b/gcc/testsuite/gfortran.dg/gomp/target-device-ancestor-4.f90
@@ -0,0 +1,14 @@
+! { dg-do compile }
+! { dg-additional-options "-fdump-tree-original" }
+
+! Test to ensure that device-modifier 'ancestor' is parsed correctly in
+! device clauses.
+
+!$omp requires reverse_offload ! { dg-error "Sorry, 'reverse_offload' clause at \\(1\\) on REQUIRES directive is not yet supported" }
+
+!$omp target device (ancestor : 1) ! { dg-message "" "sorry, unimplemented: 'ancestor' not yet supported" { xfail *-*-* } }
+!$omp end target
+
+end
+
+! { dg-final { scan-tree-dump "pragma omp target \[^\n\r)]*device\\(ancestor:1\\)" "original" } }
diff --git a/gcc/tree-pretty-print.c b/gcc/tree-pretty-print.c
index e103d2c6bd44623b81d83a0a3798204ad34fb4a0..5495942043883b15f30d1bec3470cd68ec73fdca 100644
--- a/gcc/tree-pretty-print.c
+++ b/gcc/tree-pretty-print.c
@@ -986,6 +986,8 @@ dump_omp_clause (pretty_printer *pp, tree clause, int spc, dump_flags_t flags)
case OMP_CLAUSE_DEVICE:
pp_string (pp, "device(");
+ if (OMP_CLAUSE_DEVICE_ANCESTOR (clause))
+ pp_string (pp, "ancestor:");
dump_generic_node (pp, OMP_CLAUSE_DEVICE_ID (clause),
spc, flags, false);
pp_right_paren (pp);
diff --git a/gcc/tree.h b/gcc/tree.h
index 060a41f6991f5455644fcdf00817b4b46de515e8..c32193219efce626fedbe9cc5b7860bc1c535d29 100644
--- a/gcc/tree.h
+++ b/gcc/tree.h
@@ -1691,6 +1691,10 @@ class auto_suppress_location_wrappers
#define OMP_CLAUSE_DEVICE_TYPE_KIND(NODE) \
(OMP_CLAUSE_SUBCODE_CHECK (NODE, OMP_CLAUSE_DEVICE_TYPE)->omp_clause.subcode.device_type_kind)
+/* True if there is a device clause with a device-modifier 'ancestor'. */
+#define OMP_CLAUSE_DEVICE_ANCESTOR(NODE) \
+ (OMP_CLAUSE_SUBCODE_CHECK (NODE, OMP_CLAUSE_DEVICE)->base.public_flag)
+
#define OMP_CLAUSE_COLLAPSE_EXPR(NODE) \
OMP_CLAUSE_OPERAND (OMP_CLAUSE_SUBCODE_CHECK (NODE, OMP_CLAUSE_COLLAPSE), 0)
#define OMP_CLAUSE_COLLAPSE_ITERVAR(NODE) \