[C/C++, OpenACC] Reject vars of different scope in acc declare (PR94120)

	gcc/c/
	PR middle-end/94120
	* c-decl.c (c_check_in_current_scope): New function.
	* c-tree.h (c_check_in_current_scope): Declare it.
	* c-parser.c (c_parser_oacc_declare): Add check that variables
	are declared in the same scope as the directive. Fix handling
	of namespace vars.

	gcc/cp/
	PR middle-end/94120
	* paser.c (cp_parser_oacc_declare): Add check that variables
	are declared in the same scope as the directive.

	gcc/testsuite/
	PR middle-end/94120
	* c-c++-common/goacc/declare-pr94120.c: New.
	* g++.dg/declare-pr94120.C: New.

	libgomp/testsuite/
	PR middle-end/94120
	* libgomp.oacc-c++/declare-pr94120.C: New.

10 files changed, 112 insertions, 3 deletions

diff --git a/gcc/c/ChangeLog b/gcc/c/ChangeLog
index d30d2b0..482a01b 100644
--- a/gcc/c/ChangeLog
+++ b/gcc/c/ChangeLog
@@ -1,3 +1,12 @@
+2020-04-08  Tobias Burnus  <tobias@codesourcery.com>
+
+	PR middle-end/94120
+	* c-decl.c (c_check_in_current_scope): New function.
+	* c-tree.h (c_check_in_current_scope): Declare it.
+	* c-parser.c (c_parser_oacc_declare): Add check that variables
+	are declared in the same scope as the directive. Fix handling
+	of namespace vars.
+
 2020-04-07  Jakub Jelinek  <jakub@redhat.com>
 
 	PR c++/94512
diff --git a/gcc/c/c-decl.c b/gcc/c/c-decl.c
index b31d99f..0b7f437 100644
--- a/gcc/c/c-decl.c
+++ b/gcc/c/c-decl.c
@@ -12041,4 +12041,12 @@ c_check_omp_declare_reduction_r (tree *tp, int *, void *data)
   return NULL_TREE;
 }
 
+
+bool
+c_check_in_current_scope (tree decl)
+{
+  struct c_binding *b = I_SYMBOL_BINDING (DECL_NAME (decl));
+  return b != NULL && B_IN_CURRENT_SCOPE (b);
+}
+
 #include "gt-c-c-decl.h"
diff --git a/gcc/c/c-parser.c b/gcc/c/c-parser.c
index 17a28e9..d1c954c 100644
--- a/gcc/c/c-parser.c
+++ b/gcc/c/c-parser.c
@@ -16580,6 +16580,15 @@ c_parser_oacc_declare (c_parser *parser)
 	  break;
 	}
 
+      if (!c_check_in_current_scope (decl))
+	{
+	  error_at (loc,
+		    "%qD must be a variable declared in the same scope as "
+		    "%<#pragma acc declare%>", decl);
+	  error = true;
+	  continue;
+	}
+
       if (lookup_attribute ("omp declare target", DECL_ATTRIBUTES (decl))
 	  || lookup_attribute ("omp declare target link",
 			       DECL_ATTRIBUTES (decl)))
diff --git a/gcc/c/c-tree.h b/gcc/c/c-tree.h
index 2015827..9466886 100644
--- a/gcc/c/c-tree.h
+++ b/gcc/c/c-tree.h
@@ -793,6 +793,7 @@ extern tree c_omp_reduction_id (enum tree_code, tree);
 extern tree c_omp_reduction_decl (tree);
 extern tree c_omp_reduction_lookup (tree, tree);
 extern tree c_check_omp_declare_reduction_r (tree *, int *, void *);
+extern bool c_check_in_current_scope (tree);
 extern void c_pushtag (location_t, tree, tree);
 extern void c_bind (location_t, tree, bool);
 extern bool tag_exists_p (enum tree_code, tree);
diff --git a/gcc/cp/ChangeLog b/gcc/cp/ChangeLog
index 31f1fc4..561eb4e 100644
--- a/gcc/cp/ChangeLog
+++ b/gcc/cp/ChangeLog
@@ -1,3 +1,9 @@
+2020-04-08  Tobias Burnus  <tobias@codesourcery.com>
+
+	PR middle-end/94120
+	* paser.c (cp_parser_oacc_declare): Add check that variables
+	are declared in the same scope as the directive.
+
 2020-04-07  Jason Merrill  <jason@redhat.com>
 
 	PR c++/94480
diff --git a/gcc/cp/parser.c b/gcc/cp/parser.c
index a95d431..fec5203 100644
--- a/gcc/cp/parser.c
+++ b/gcc/cp/parser.c
@@ -40906,6 +40906,7 @@ cp_parser_oacc_declare (cp_parser *parser, cp_token *pragma_tok)
 {
   tree clauses, stmt;
   bool error = false;
+  bool found_in_scope = global_bindings_p ();
 
   clauses = cp_parser_oacc_all_clauses (parser, OACC_DECLARE_CLAUSE_MASK,
 					"#pragma acc declare", pragma_tok, true);
@@ -40978,6 +40979,22 @@ cp_parser_oacc_declare (cp_parser *parser, cp_token *pragma_tok)
 	  break;
 	}
 
+      if (!found_in_scope)
+	for (tree d = current_binding_level->names; d; d = TREE_CHAIN (d))
+	  if (d == decl)
+	    {
+	      found_in_scope = true;
+	      break;
+	    }
+      if (!found_in_scope)
+	{
+	  error_at (loc,
+		    "%qD must be a variable declared in the same scope as "
+		    "%<#pragma acc declare%>", decl);
+	  error = true;
+	  continue;
+	}
+
       if (lookup_attribute ("omp declare target", DECL_ATTRIBUTES (decl))
 	  || lookup_attribute ("omp declare target link",
 			       DECL_ATTRIBUTES (decl)))
@@ -40999,7 +41016,7 @@ cp_parser_oacc_declare (cp_parser *parser, cp_token *pragma_tok)
 
 	  DECL_ATTRIBUTES (decl)
 	    = tree_cons (id, NULL_TREE, DECL_ATTRIBUTES (decl));
-	  if (global_bindings_p ())
+	  if (current_binding_level->kind == sk_namespace)
 	    {
 	      symtab_node *node = symtab_node::get (decl);
 	      if (node != NULL)
@@ -41016,7 +41033,7 @@ cp_parser_oacc_declare (cp_parser *parser, cp_token *pragma_tok)
 	}
     }
 
-  if (error || global_bindings_p ())
+  if (error || current_binding_level->kind == sk_namespace)
     return NULL_TREE;
 
   stmt = make_node (OACC_DECLARE);
diff --git a/gcc/omp-grid.c b/gcc/omp-grid.c
index b98e45d..ba635fd 100644
--- a/gcc/omp-grid.c
+++ b/gcc/omp-grid.c
@@ -1065,7 +1065,7 @@ grid_eliminate_combined_simd_part (gomp_for *parloop)
   while (*pc)
     {
       tree c = *pc;
-      switch (TREE_CODE (c))
+      switch (OMP_CLAUSE_CODE (c))
 	{
 	case OMP_CLAUSE_LINEAR:
 	  {
diff --git a/gcc/testsuite/ChangeLog b/gcc/testsuite/ChangeLog
index 3cbf891..e4e6ecf 100644
--- a/gcc/testsuite/ChangeLog
+++ b/gcc/testsuite/ChangeLog
@@ -1,3 +1,9 @@
+2020-04-08  Tobias Burnus  <tobias@codesourcery.com>
+
+	PR middle-end/94120
+	* c-c++-common/goacc/declare-pr94120.c: New.
+	* g++.dg/declare-pr94120.C: New.
+
 2020-04-07  Jakub Jelinek  <jakub@redhat.com>
 
 	PR rtl-optimization/94291
diff --git a/gcc/testsuite/c-c++-common/goacc/declare-pr94120.c b/gcc/testsuite/c-c++-common/goacc/declare-pr94120.c
new file mode 100644
index 0000000..21b2cc1
--- /dev/null
+++ b/gcc/testsuite/c-c++-common/goacc/declare-pr94120.c
@@ -0,0 +1,23 @@
+/* { dg-do compile }  */
+
+/* PR middle-end/94120  */
+
+void foo()
+{
+  int foo;
+  {
+    #pragma acc declare copy(foo)  /* { dg-error "'foo' must be a variable declared in the same scope as '#pragma acc declare'" }  */
+  }
+}
+
+void
+f_data (void)
+{
+  int B[10];
+#pragma acc data
+  {
+# pragma acc declare copy(B)  /* { dg-error "'B' must be a variable declared in the same scope as '#pragma acc declare'" }  */
+    for (int i = 0; i < 10; i++)
+      B[i] = -i;
+  }
+}
diff --git a/gcc/testsuite/g++.dg/declare-pr94120.C b/gcc/testsuite/g++.dg/declare-pr94120.C
new file mode 100644
index 0000000..8515c4f
--- /dev/null
+++ b/gcc/testsuite/g++.dg/declare-pr94120.C
@@ -0,0 +1,30 @@
+/* { dg-do compile }  */
+
+/* PR middle-end/94120  */
+
+int b[8];
+#pragma acc declare create (b)
+ 
+namespace my {
+ int d[8] = { 1, 2, 3, 4, 5, 6, 7, 8 };
+ #pragma acc declare copyin (d)
+};
+
+namespace outer {
+  namespace inner {
+    int e[8] = { 1, 2, 3, 4, 5, 6, 7, 8 };
+    #pragma acc declare copyin (e)
+  };
+};
+
+int f[8] = { 1, 2, 3, 4, 5, 6, 7, 8 };
+namespace my {
+ #pragma acc declare copyin (f)
+};
+
+namespace outer {
+  int g[8] = { 1, 2, 3, 4, 5, 6, 7, 8 };
+  namespace inner {
+    #pragma acc declare copyin (g)
+  };
+};