From: Thomas Schwinge <thomas@codesourcery.com>
Date: Wed, 4 Mar 2020 16:58:33 +0000 (+0100)
Subject: Handle 'omp declare target' attribute set for both OpenACC and OpenMP 'target' [PR894... 
X-Git-Url: https://git.libre-soc.org/?a=commitdiff_plain;h=ff3f862b451496dd4afbe2dbfae82afab59a42c6;p=gcc.git

Handle 'omp declare target' attribute set for both OpenACC and OpenMP 'target' [PR89433, PR93465]

... which as of PR89433 commit b48f44bf77a39fefc238a16cf1225c6464c82406 causes
an ICE.  Not sure if this is actually supposed to be valid or invalid code.
Until the interactions between OpenACC and OpenMP 'target' get defined
properly, make this a compile-time error.

	gcc/
	PR middle-end/89433
	PR middle-end/93465
	* omp-general.c (oacc_verify_routine_clauses): Diagnose if
	"#pragma omp declare target" has also been applied.
	gcc/testsuite/
	PR middle-end/89433
	PR middle-end/93465
	* c-c++-common/goacc-gomp/pr93465-1.c: New file.
---

diff --git a/gcc/ChangeLog b/gcc/ChangeLog
index bce700e472e..33e980bd977 100644
--- a/gcc/ChangeLog
+++ b/gcc/ChangeLog
@@ -1,3 +1,10 @@
+2020-04-10  Thomas Schwinge  <thomas@codesourcery.com>
+
+	PR middle-end/89433
+	PR middle-end/93465
+	* omp-general.c (oacc_verify_routine_clauses): Diagnose if
+	"#pragma omp declare target" has also been applied.
+
 2020-04-09  Jozef Lawrynowicz  <jozef.l@mittosystems.com>
 
 	* config/msp430/msp430.c (msp430_expand_epilogue): Use emit_jump_insn
diff --git a/gcc/omp-general.c b/gcc/omp-general.c
index f107f4c050f..49023f42c47 100644
--- a/gcc/omp-general.c
+++ b/gcc/omp-general.c
@@ -1776,6 +1776,19 @@ oacc_verify_routine_clauses (tree fndecl, tree *clauses, location_t loc,
     = lookup_attribute ("omp declare target", DECL_ATTRIBUTES (fndecl));
   if (attr != NULL_TREE)
     {
+      /* Diagnose if "#pragma omp declare target" has also been applied.  */
+      if (TREE_VALUE (attr) == NULL_TREE)
+	{
+	  /* See <https://gcc.gnu.org/PR93465>; the semantics of combining
+	     OpenACC and OpenMP 'target' are not clear.  */
+	  error_at (loc,
+		    "cannot apply %<%s%> to %qD, which has also been"
+		    " marked with an OpenMP 'declare target' directive",
+		    routine_str, fndecl);
+	  /* Incompatible.  */
+	  return -1;
+	}
+
       /* If a "#pragma acc routine" has already been applied, just verify
 	 this one for compatibility.  */
       /* Collect previous directive's clauses.  */
diff --git a/gcc/testsuite/ChangeLog b/gcc/testsuite/ChangeLog
index bced3e27d86..235d481e6cf 100644
--- a/gcc/testsuite/ChangeLog
+++ b/gcc/testsuite/ChangeLog
@@ -1,3 +1,9 @@
+2020-04-10  Thomas Schwinge  <thomas@codesourcery.com>
+
+	PR middle-end/89433
+	PR middle-end/93465
+	* c-c++-common/goacc-gomp/pr93465-1.c: New file.
+
 2020-04-10  Iain Buclaw  <ibuclaw@gdcproject.org>
 
 	* lib/gdc.exp (gdc_link_flags): Remove libdruntime library paths.
diff --git a/gcc/testsuite/c-c++-common/goacc-gomp/pr93465-1.c b/gcc/testsuite/c-c++-common/goacc-gomp/pr93465-1.c
new file mode 100644
index 00000000000..c8b9135d997
--- /dev/null
+++ b/gcc/testsuite/c-c++-common/goacc-gomp/pr93465-1.c
@@ -0,0 +1,56 @@
+#pragma omp declare target
+#pragma acc routine seq /* { dg-error "cannot apply '#pragma acc routine' to '\(void \)?f1\(\\(\\)\)?', which has also been marked with an OpenMP 'declare target' directive" } */
+void f1 (void) {}
+#pragma omp end declare target
+
+#pragma omp declare target
+void f1 (void);
+
+#pragma acc routine seq /* { dg-error "cannot apply '#pragma acc routine' to '\(void \)?f1\(\\(\\)\)?', which has also been marked with an OpenMP 'declare target' directive" } */
+void f1 (void);
+
+
+
+#pragma omp declare target
+#pragma acc routine /* { dg-error "cannot apply '#pragma acc routine' to '\(void \)?f2\(\\(\\)\)?', which has also been marked with an OpenMP 'declare target' directive" } */
+extern void f2 (void);
+#pragma omp end declare target
+
+#pragma omp declare target
+extern void f2 (void);
+#pragma omp end declare target
+
+#pragma acc routine gang /* { dg-error "cannot apply '#pragma acc routine' to '\(void \)?f2\(\\(\\)\)?', which has also been marked with an OpenMP 'declare target' directive" } */
+extern void f2 (void);
+
+
+#pragma omp declare target
+#pragma acc routine gang /* { dg-error "cannot apply '#pragma acc routine' to '\(void \)?f3\(\\(\\)\)?', which has also been marked with an OpenMP 'declare target' directive" } */
+void f3 (void);
+#pragma omp end declare target
+
+#pragma omp declare target
+void f3 (void) {}
+#pragma omp end declare target
+
+#pragma acc routine (f3) gang /* { dg-error "cannot apply '#pragma acc routine' to '\(void \)?f3\(\\(\\)\)?', which has also been marked with an OpenMP 'declare target' directive" } */
+
+
+/* Surprisingly, this diagnosis also works for '#pragma acc routine' first,
+   followed by '#pragma omp declare target'; the latter gets applied first.  */
+
+
+#pragma acc routine /* { dg-error "cannot apply '#pragma acc routine' to '\(void \)?f4\(\\(\\)\)?', which has also been marked with an OpenMP 'declare target' directive" } */
+extern void f4 (void);
+
+#pragma omp declare target
+extern void f4 (void);
+#pragma omp end declare target
+
+
+#pragma acc routine gang /* { dg-error "cannot apply '#pragma acc routine' to '\(void \)?f5\(\\(\\)\)?', which has also been marked with an OpenMP 'declare target' directive" } */
+void f5 (void) {}
+
+#pragma omp declare target
+extern void f5 (void);
+#pragma omp end declare target