From: Thomas Schwinge <thomas@codesourcery.com>
Date: Wed, 25 Nov 2020 12:03:52 +0000 (+0100)
Subject: Fix templatized C++ OpenACC 'cache' directive ICEs
X-Git-Url: https://git.libre-soc.org/?a=commitdiff_plain;h=0cab70604cfda30bc64351b39493ef884ff7ba10;p=gcc.git

Fix templatized C++ OpenACC 'cache' directive ICEs

This has been broken forever, whoops...

	gcc/cp/
	* pt.c (tsubst_omp_clauses): Handle 'OMP_CLAUSE__CACHE_'.
	(tsubst_expr): Handle 'OACC_CACHE'.
	gcc/testsuite/
	* c-c++-common/goacc/cache-1.c: Update.
	* c-c++-common/goacc/cache-2.c: Likewise.
	* g++.dg/goacc/cache-1.C: New.
	* g++.dg/goacc/cache-2.C: Likewise.
	libgomp/
	* testsuite/libgomp.oacc-c++/cache-1.C: New.
	* testsuite/libgomp.oacc-c-c++-common/cache-1.c: Update.
---

diff --git a/gcc/cp/pt.c b/gcc/cp/pt.c
index fdd7f2d457b..4fb0bc82c31 100644
--- a/gcc/cp/pt.c
+++ b/gcc/cp/pt.c
@@ -17245,6 +17245,7 @@ tsubst_omp_clauses (tree clauses, enum c_omp_region_type ort,
 	case OMP_CLAUSE_FROM:
 	case OMP_CLAUSE_TO:
 	case OMP_CLAUSE_MAP:
+	case OMP_CLAUSE__CACHE_:
 	case OMP_CLAUSE_NONTEMPORAL:
 	case OMP_CLAUSE_USE_DEVICE_PTR:
 	case OMP_CLAUSE_USE_DEVICE_ADDR:
@@ -18761,6 +18762,7 @@ tsubst_expr (tree t, tree args, tsubst_flags_t complain, tree in_decl,
       add_stmt (t);
       break;
 
+    case OACC_CACHE:
     case OACC_ENTER_DATA:
     case OACC_EXIT_DATA:
     case OACC_UPDATE:
diff --git a/gcc/testsuite/c-c++-common/goacc/cache-1.c b/gcc/testsuite/c-c++-common/goacc/cache-1.c
index 1d4759e738c..242f3c612fd 100644
--- a/gcc/testsuite/c-c++-common/goacc/cache-1.c
+++ b/gcc/testsuite/c-c++-common/goacc/cache-1.c
@@ -1,9 +1,15 @@
-/* OpenACC cache directive: valid usage.  */
-/* For execution testing, this file is "#include"d from
-   libgomp/testsuite/libgomp.oacc-c-c++-common/cache-1.c.  */
+/* OpenACC 'cache' directive: valid usage.  */
 
-int
-main (int argc, char **argv)
+/* See also corresponding C++ variant: '../../g++.dg/goacc/cache-1.C'.  */
+
+/* For execution testing, this file is '#include'd from
+   '../../../../libgomp/testsuite/libgomp.oacc-c-c++-common/cache-1.c'.  */
+
+#ifdef TEMPLATIZE
+template <int N>
+#endif
+static void
+test ()
 {
 #define N   2
     int a[N], b[N];
@@ -61,6 +67,4 @@ main (int argc, char **argv)
         if (a[i] != b[i])
             __builtin_abort ();
     }
-
-    return 0;
 }
diff --git a/gcc/testsuite/c-c++-common/goacc/cache-2.c b/gcc/testsuite/c-c++-common/goacc/cache-2.c
index d1181d1b6e7..80b925e5112 100644
--- a/gcc/testsuite/c-c++-common/goacc/cache-2.c
+++ b/gcc/testsuite/c-c++-common/goacc/cache-2.c
@@ -1,7 +1,9 @@
-/* OpenACC cache directive: invalid usage.  */
+/* OpenACC 'cache' directive: invalid usage.  */
 
-int
-main (int argc, char **argv)
+/* See also corresponding C++ variant: '../../g++.dg/goacc/cache-2.C'.  */
+
+static void
+test ()
 {
 #define N   2
     int a[N], b[N];
@@ -52,6 +54,4 @@ main (int argc, char **argv)
         if (a[i] != b[i])
             __builtin_abort ();
     }
-
-    return 0;
 }
diff --git a/gcc/testsuite/g++.dg/goacc/cache-1.C b/gcc/testsuite/g++.dg/goacc/cache-1.C
new file mode 100644
index 00000000000..a8d5ab32016
--- /dev/null
+++ b/gcc/testsuite/g++.dg/goacc/cache-1.C
@@ -0,0 +1,15 @@
+/* OpenACC 'cache' directive: valid usage.  */
+
+/* See also corresponding C/C++ variant '../../c-c++-common/goacc/cache-1.c'.  */
+
+/* For execution testing, this file is '#include'd from
+   '../../../../libgomp/testsuite/libgomp.oacc-c++/cache-1.C'.  */
+
+#define TEMPLATIZE
+#include "../../c-c++-common/goacc/cache-1.c"
+
+static void
+instantiate ()
+{
+  &test<0>;
+}
diff --git a/gcc/testsuite/g++.dg/goacc/cache-2.C b/gcc/testsuite/g++.dg/goacc/cache-2.C
new file mode 100644
index 00000000000..ef0b8a9e76f
--- /dev/null
+++ b/gcc/testsuite/g++.dg/goacc/cache-2.C
@@ -0,0 +1,64 @@
+/* OpenACC 'cache' directive: invalid usage.  */
+
+/* See also corresponding C/C++ variant '../../c-c++-common/goacc/cache-2.c'.  */
+
+template <int N>
+static void
+test ()
+{
+#define N   2
+    int a[N], b[N];
+    int i;
+
+    for (i = 0; i < N; i++)
+    {
+        a[i] = 3;
+        b[i] = 0;
+    }
+
+#pragma acc parallel copyin (a[0:N]) copyout (b[0:N])
+{
+    int ii;
+
+    for (ii = 0; ii < N; ii++)
+    {
+        const int idx = ii;
+        int n = 1;
+        const int len = n;
+
+#pragma acc cache /* { dg-error "expected '\\\(' before end of line" } */
+#pragma acc cache a[0:N] /* { dg-error "expected '\\\(' before 'a'" } */
+	/* { dg-bogus "expected end of line before 'a'" "" { xfail c++ } .-1 } */
+#pragma acc cache (a) /* { dg-error "expected '\\\['" } */
+#pragma acc cache ( /* { dg-error "expected (identifier|unqualified-id) before end of line" } */
+#pragma acc cache () /* { dg-error "expected (identifier|unqualified-id) before '\\\)' token" } */
+#pragma acc cache (,) /* { dg-error "expected (identifier|unqualified-id) before '(,|\\\))' token" } */
+#pragma acc cache (a[0:N] /* { dg-error "expected '\\\)' before end of line" } */
+#pragma acc cache (a[0:N],) /* { dg-error "expected (identifier|unqualified-id) before '(,|\\\))' token" "" { xfail c } } */
+#pragma acc cache (a[0:N]) copyin (a[0:N]) /* { dg-error "expected end of line before 'copyin'" } */
+#pragma acc cache () /* { dg-error "expected (identifier|unqualified-id) before '\\\)' token" } */
+#pragma acc cache (a[0:N] b[0:N]) /* { dg-error "expected '\\\)' before 'b'" } */
+#pragma acc cache (a[0:N] b[0:N}) /* { dg-error "expected '\\\)' before 'b'" } */
+	/* { dg-bogus "expected end of line before '\\\}' token" "" { xfail c++ } .-1 } */
+#pragma acc cache (a[0:N] /* { dg-error "expected '\\\)' before end of line" } */
+#pragma acc cache (a[0:N]) ( /* { dg-error "expected end of line before '\\(' token" } */
+#pragma acc cache (a[0:N]) ii /* { dg-error "expected end of line before 'ii'" } */
+#pragma acc cache (a[0:N] ii) /* { dg-error "expected '\\)' before 'ii'" } */
+
+        b[ii] = a[ii];
+    }
+}
+
+
+    for (i = 0; i < N; i++)
+    {
+        if (a[i] != b[i])
+            __builtin_abort ();
+    }
+}
+
+static void
+instantiate ()
+{
+  &test<0>;
+}
diff --git a/libgomp/testsuite/libgomp.oacc-c++/cache-1.C b/libgomp/testsuite/libgomp.oacc-c++/cache-1.C
new file mode 100644
index 00000000000..fcb1f84eef9
--- /dev/null
+++ b/libgomp/testsuite/libgomp.oacc-c++/cache-1.C
@@ -0,0 +1,13 @@
+/* OpenACC 'cache' directive.  */
+
+/* See also corresponding C/C++ variant '../libgomp.oacc-c-c++-common/cache-1.c'.  */
+
+#include "../../../gcc/testsuite/g++.dg/goacc/cache-1.C"
+
+int
+main (int argc, char *argv[])
+{
+  test<0> ();
+
+  return 0;
+}
diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/cache-1.c b/libgomp/testsuite/libgomp.oacc-c-c++-common/cache-1.c
index 16aaed5df70..c0dddb3c4f3 100644
--- a/libgomp/testsuite/libgomp.oacc-c-c++-common/cache-1.c
+++ b/libgomp/testsuite/libgomp.oacc-c-c++-common/cache-1.c
@@ -1,3 +1,13 @@
-/* OpenACC cache directive.  */
+/* OpenACC 'cache' directive.  */
+
+/* See also corresponding C++ variant '../libgomp.oacc-c++/cache-1.C'.  */
 
 #include "../../../gcc/testsuite/c-c++-common/goacc/cache-1.c"
+
+int
+main (int argc, char *argv[])
+{
+  test ();
+
+  return 0;
+}