https://gcc.gnu.org/g:23416673c4920bd387b74d6f8aed05e1e85d60f7
commit 23416673c4920bd387b74d6f8aed05e1e85d60f7
Author: Sandra Loosemore <sloosem...@baylibre.com>
Date: Wed May 15 14:34:16 2024 +0200
OpenMP: Update "declare target"/OpenMP context interaction
The code and test case previously implemented the OpenMP 5.0 spec,
which said in section 2.3.1:
"For functions within a declare target block, the target trait is added
to the beginning of the set..."
In OpenMP 5.1, this was changed to
"For device routines, the target trait is added to the beginning of
the set..."
In OpenMP 5.2 and TR12, it says:
"For procedures that are determined to be target function variants
by a declare target directive..."
The definition of "device routine" in OpenMP 5.1 is confusing, but
certainly the intent of the later versions of the spec is clear that
it doesn't just apply to functions within a begin declare target/end
declare target block.
The only use of the "omp declare target block" function attribute was
to support the 5.0 language, so it can be removed. This patch changes
the context augmentation to use the "omp declare target" attribute
instead.
gcc/c-family/ChangeLog
* c-attribs.cc (c_common_gnu_attributes): Delete "omp declare
target block".
gcc/c/ChangeLog
* c-decl.cc (c_decl_attributes): Don't add "omp declare target
block".
gcc/cp/decl2.cc
* decl2.cc (cplus_decl_attributes): Don't add "omp declare target
block".
gcc/ChangeLog
* omp-general.cc (omp_complete_construct_context): Check
"omp declare target" attribute, not "omp declare target block".
gcc/testsuite/ChangeLog
* c-c++-common/gomp/declare-target-indirect-2.c : Adjust
expected output for removal of "omp declare target block".
* c-c++-common/gomp/declare-variant-8.c: Likewise, the variant
call to f20 is now resolved differently.
* c-c++-common/gomp/reverse-offload-1.c: Adjust expected output.
* gfortran.dg/gomp/declare-variant-8.f90: Likewise, both f18
and f20 now resolve to the variant. Delete obsolete comments.
Diff:
---
gcc/ChangeLog.omp | 5 +++++
gcc/c-family/ChangeLog.omp | 5 +++++
gcc/c-family/c-attribs.cc | 2 --
gcc/c/ChangeLog.omp | 5 +++++
gcc/c/c-decl.cc | 8 ++------
gcc/cp/ChangeLog.omp | 5 +++++
gcc/cp/decl2.cc | 9 ++-------
gcc/omp-general.cc | 2 +-
gcc/testsuite/ChangeLog.omp | 10 ++++++++++
gcc/testsuite/c-c++-common/gomp/declare-target-indirect-2.c | 10 +++++-----
gcc/testsuite/c-c++-common/gomp/declare-variant-8.c | 4 ++--
gcc/testsuite/c-c++-common/gomp/reverse-offload-1.c | 2 +-
gcc/testsuite/gfortran.dg/gomp/declare-variant-8.f90 | 12 ++----------
13 files changed, 45 insertions(+), 34 deletions(-)
diff --git a/gcc/ChangeLog.omp b/gcc/ChangeLog.omp
index b0f17237a79..e95ec6a43e2 100644
--- a/gcc/ChangeLog.omp
+++ b/gcc/ChangeLog.omp
@@ -1,3 +1,8 @@
+2024-05-04 Sandra Loosemore <sloosem...@baylibre.com>
+
+ * omp-general.cc (omp_complete_construct_context): Check
+ "omp declare target" attribute, not "omp declare target block".
+
2024-05-04 Sandra Loosemore <sloosem...@baylibre.com>
* cgraph.cc (symbol_table::create_edge): Don't set
diff --git a/gcc/c-family/ChangeLog.omp b/gcc/c-family/ChangeLog.omp
index 543bba589af..e23588dde7d 100644
--- a/gcc/c-family/ChangeLog.omp
+++ b/gcc/c-family/ChangeLog.omp
@@ -1,3 +1,8 @@
+2024-05-04 Sandra Loosemore <sloosem...@baylibre.com>
+
+ * c-attribs.cc (c_common_gnu_attributes): Delete "omp declare
+ target block".
+
2024-05-04 Sandra Loosemore <sloosem...@baylibre.com>
* c-common.h (enum c_omp_directive_kind): Add C_OMP_DIR_META.
diff --git a/gcc/c-family/c-attribs.cc b/gcc/c-family/c-attribs.cc
index 04e39b41bdf..582d99ada1b 100644
--- a/gcc/c-family/c-attribs.cc
+++ b/gcc/c-family/c-attribs.cc
@@ -570,8 +570,6 @@ const struct attribute_spec c_common_gnu_attributes[] =
handle_omp_declare_target_attribute, NULL },
{ "omp declare target nohost", 0, 0, true, false, false, false,
handle_omp_declare_target_attribute, NULL },
- { "omp declare target block", 0, 0, true, false, false, false,
- handle_omp_declare_target_attribute, NULL },
{ "non overlapping", 0, 0, true, false, false, false,
handle_non_overlapping_attribute, NULL },
{ "alloc_align", 1, 1, false, true, true, false,
diff --git a/gcc/c/ChangeLog.omp b/gcc/c/ChangeLog.omp
index b246b47e9f6..5221390d138 100644
--- a/gcc/c/ChangeLog.omp
+++ b/gcc/c/ChangeLog.omp
@@ -1,3 +1,8 @@
+2024-05-04 Sandra Loosemore <sloosem...@baylibre.com>
+
+ * c-decl.cc (c_decl_attributes): Don't add "omp declare target
+ block".
+
2024-05-04 Sandra Loosemore <sloosem...@baylibre.com>
* c-parser.c (c_parser_omp_context_selector): Remove metadirective_p
diff --git a/gcc/c/c-decl.cc b/gcc/c/c-decl.cc
index e4540fe3e65..3aed9f50dd2 100644
--- a/gcc/c/c-decl.cc
+++ b/gcc/c/c-decl.cc
@@ -5414,12 +5414,8 @@ c_decl_attributes (tree *node, tree attributes, int
flags)
attributes = tree_cons (get_identifier ("omp declare target implicit"),
NULL_TREE, attributes);
else
- {
- attributes = tree_cons (get_identifier ("omp declare target"),
- NULL_TREE, attributes);
- attributes = tree_cons (get_identifier ("omp declare target block"),
- NULL_TREE, attributes);
- }
+ attributes = tree_cons (get_identifier ("omp declare target"),
+ NULL_TREE, attributes);
if (TREE_CODE (*node) == FUNCTION_DECL)
{
int device_type
diff --git a/gcc/cp/ChangeLog.omp b/gcc/cp/ChangeLog.omp
index 88ac4bd975d..a8f9285a420 100644
--- a/gcc/cp/ChangeLog.omp
+++ b/gcc/cp/ChangeLog.omp
@@ -1,3 +1,8 @@
+2024-05-04 Sandra Loosemore <sloosem...@baylibre.com>
+
+ * decl2.cc (cplus_decl_attributes): Don't add "omp declare target
+ block".
+
2024-05-04 Sandra Loosemore <sloosem...@baylibre.com>
* cp-tree.h (struct saved_scope): Add new field
diff --git a/gcc/cp/decl2.cc b/gcc/cp/decl2.cc
index 9048098ec5f..7ecca8180fc 100644
--- a/gcc/cp/decl2.cc
+++ b/gcc/cp/decl2.cc
@@ -1784,13 +1784,8 @@ cplus_decl_attributes (tree *decl, tree attributes, int
flags)
= tree_cons (get_identifier ("omp declare target implicit"),
NULL_TREE, attributes);
else
- {
- attributes = tree_cons (get_identifier ("omp declare target"),
- NULL_TREE, attributes);
- attributes
- = tree_cons (get_identifier ("omp declare target block"),
- NULL_TREE, attributes);
- }
+ attributes = tree_cons (get_identifier ("omp declare target"),
+ NULL_TREE, attributes);
if (TREE_CODE (*decl) == FUNCTION_DECL)
{
cp_omp_declare_target_attr &last
diff --git a/gcc/omp-general.cc b/gcc/omp-general.cc
index f181a4bbc46..361dcb0c46d 100644
--- a/gcc/omp-general.cc
+++ b/gcc/omp-general.cc
@@ -2658,7 +2658,7 @@ omp_complete_construct_context (tree construct_context,
bool *completep)
}
/* Add target trait when in a target variant. */
- if (lookup_attribute ("omp declare target block", attributes))
+ if (lookup_attribute ("omp declare target", attributes))
construct_context = make_trait_selector (OMP_TRAIT_CONSTRUCT_TARGET,
NULL_TREE, NULL_TREE,
construct_context);
diff --git a/gcc/testsuite/ChangeLog.omp b/gcc/testsuite/ChangeLog.omp
index e602028629f..5d1528e4ba8 100644
--- a/gcc/testsuite/ChangeLog.omp
+++ b/gcc/testsuite/ChangeLog.omp
@@ -1,3 +1,13 @@
+2024-05-04 Sandra Loosemore <sloosem...@baylibre.com>
+
+ * c-c++-common/gomp/declare-target-indirect-2.c : Adjust
+ expected output for removal of "omp declare target block".
+ * c-c++-common/gomp/declare-variant-8.c: Likewise, the variant
+ call to f20 is now resolved differently.
+ * c-c++-common/gomp/reverse-offload-1.c: Adjust expected output.
+ * gfortran.dg/gomp/declare-variant-8.f90: Likewise, both f18
+ and f20 now resolve to the variant. Delete obsolete comments.
+
2024-05-04 Sandra Loosemore <sloosem...@baylibre.com>
* c-c++-common/gomp/declare-variant-12.c: Adjust expected behavior.
diff --git a/gcc/testsuite/c-c++-common/gomp/declare-target-indirect-2.c
b/gcc/testsuite/c-c++-common/gomp/declare-target-indirect-2.c
index 6ba278b3ef0..75a205feb95 100644
--- a/gcc/testsuite/c-c++-common/gomp/declare-target-indirect-2.c
+++ b/gcc/testsuite/c-c++-common/gomp/declare-target-indirect-2.c
@@ -4,12 +4,12 @@
#pragma omp begin declare target indirect
void fn1 (void) { }
#pragma omp end declare target
-/* { dg-final { scan-tree-dump "__attribute__\\\(\\\(omp declare target, omp
declare target block, omp declare target indirect\\\)\\\)\\\nvoid fn1" "gimple"
} } */
+/* { dg-final { scan-tree-dump "__attribute__\\\(\\\(omp declare target, omp
declare target indirect\\\)\\\)\\\nvoid fn1" "gimple" } } */
#pragma omp begin declare target indirect (0)
void fn2 (void) { }
#pragma omp end declare target
-/* { dg-final { scan-tree-dump "__attribute__\\\(\\\(omp declare target, omp
declare target block\\\)\\\)\\\nvoid fn2" "gimple" } } */
+/* { dg-final { scan-tree-dump "__attribute__\\\(\\\(omp declare
target\\\)\\\)\\\nvoid fn2" "gimple" } } */
void fn3 (void) { }
#pragma omp declare target indirect to (fn3)
@@ -27,6 +27,6 @@ void fn4 (void) { }
#pragma omp declare target indirect enter(baz)
#pragma omp end declare target
#pragma omp end declare target
-/* { dg-final { scan-tree-dump "__attribute__\\\(\\\(omp declare target, omp
declare target block, omp declare target indirect\\\)\\\)\\\nint foo" "gimple"
} } */
-/* { dg-final { scan-tree-dump "__attribute__\\\(\\\(omp declare target, omp
declare target block\\\)\\\)\\\nint bar" "gimple" } } */
-/* { dg-final { scan-tree-dump "__attribute__\\\(\\\(omp declare target
indirect, omp declare target, omp declare target block\\\)\\\)\\\nint baz"
"gimple" } } */
+/* { dg-final { scan-tree-dump "__attribute__\\\(\\\(omp declare target, omp
declare target indirect\\\)\\\)\\\nint foo" "gimple" } } */
+/* { dg-final { scan-tree-dump "__attribute__\\\(\\\(omp declare
target\\\)\\\)\\\nint bar" "gimple" } } */
+/* { dg-final { scan-tree-dump "__attribute__\\\(\\\(omp declare target
indirect, omp declare target\\\)\\\)\\\nint baz" "gimple" } } */
diff --git a/gcc/testsuite/c-c++-common/gomp/declare-variant-8.c
b/gcc/testsuite/c-c++-common/gomp/declare-variant-8.c
index a7a3ba41b97..9cd706e896f 100644
--- a/gcc/testsuite/c-c++-common/gomp/declare-variant-8.c
+++ b/gcc/testsuite/c-c++-common/gomp/declare-variant-8.c
@@ -1,4 +1,4 @@
-/* { dg-do compile { target c } } */
+/* { dg-do compile } */
/* { dg-additional-options "-fdump-tree-gimple" } */
void f01 (void);
@@ -102,7 +102,7 @@ void
test3 (void)
{
#pragma omp parallel
- f20 (); /* { dg-final { scan-tree-dump-times "f20 \\\(\\\);" 1 "gimple"
} } */
+ f20 (); /* { dg-final { scan-tree-dump-times "f19 \\\(\\\);" 1 "gimple"
} } */
}
void
diff --git a/gcc/testsuite/c-c++-common/gomp/reverse-offload-1.c
b/gcc/testsuite/c-c++-common/gomp/reverse-offload-1.c
index 9a3fa5230f8..6abaddcd5f4 100644
--- a/gcc/testsuite/c-c++-common/gomp/reverse-offload-1.c
+++ b/gcc/testsuite/c-c++-common/gomp/reverse-offload-1.c
@@ -4,7 +4,7 @@
/* { dg-final { scan-tree-dump-times "__attribute__\\(\\(omp declare
target\\)\\)\[\n\r\]*int called_in_target1" 1 "omplower" } } */
/* { dg-final { scan-tree-dump-times "__attribute__\\(\\(omp declare
target\\)\\)\[\n\r\]*int called_in_target2" 1 "omplower" } } */
-/* { dg-final { scan-tree-dump-times "__attribute__\\(\\(omp declare target,
omp declare target block\\)\\)\[\n\r\]*void tg_fn" 1 "omplower" } } */
+/* { dg-final { scan-tree-dump-times "__attribute__\\(\\(omp declare
target\\)\\)\[\n\r\]*void tg_fn" 1 "omplower" } } */
/* { dg-prune-output "'reverse_offload' clause on 'requires' directive not
supported yet" } */
diff --git a/gcc/testsuite/gfortran.dg/gomp/declare-variant-8.f90
b/gcc/testsuite/gfortran.dg/gomp/declare-variant-8.f90
index d69e552eeb7..e3935768bc4 100644
--- a/gcc/testsuite/gfortran.dg/gomp/declare-variant-8.f90
+++ b/gcc/testsuite/gfortran.dg/gomp/declare-variant-8.f90
@@ -167,23 +167,15 @@ contains
end subroutine
subroutine test2 ()
- ! OpenMP 5.0 specifies that the 'target' trait should be added for
- ! functions within a declare target block, but Fortran does not have
- ! the notion of a declare target _block_, so the variant is not used here.
- ! This may change in later versions of OpenMP.
-
!$omp declare target
!$omp parallel
- call f18 () ! { dg-final { scan-tree-dump-times "f18 \\\(\\\);" 1
"gimple" } }
+ call f18 () ! { dg-final { scan-tree-dump-times "f17 \\\(\\\);" 1
"gimple" } }
!$omp end parallel
end subroutine
subroutine test3 ()
- ! In the C version, this test was used to check that the
- ! 'declare target to' form of the directive did not result in the variant
- ! being used.
!$omp parallel
- call f20 () ! { dg-final { scan-tree-dump-times "f20 \\\(\\\);" 1
"gimple" } }
+ call f20 () ! { dg-final { scan-tree-dump-times "f19 \\\(\\\);" 1
"gimple" } }
!$omp end parallel
end subroutine
