[gomp4] Fix use of declare'd vars by routine procedures.

[James Norris <jnorris at codesourcery dot com>]

I've backported this patch from trunk to gomp-4_0-branch. This patch
updates a previous patch to gomp4 in dealing with variables used
within a routine procedure.

Reference:

	https://gcc.gnu.org/ml/gcc-patches/2016-01/msg01231.html

Thanks,
Jim


==== ChangeLog entries...

        gcc/
        * gimplify.c (omp_notice_variable): Add usage check.

        gcc/testsuite/
        * c-c++-common/goacc/routine-5.c: Add tests.

        gcc/c/
        * c-typeck.c (build_external_ref): Remove usage check.

        gcc/cp/
        * semantics.c (finish_id_expression): Remove usage check.

Index: gcc/ChangeLog.gomp
===================================================================
--- gcc/ChangeLog.gomp	(revision 233104)
+++ gcc/ChangeLog.gomp	(working copy)
@@ -1,3 +1,9 @@
+2016-02-03  James Norris  <jnorris@codesourcery.com
+
+	Backport form trunk:
+	2016-02-02  James Norris  <jnorris@codesourcery.com>
+	* gimplify.c (omp_notice_variable): Add usage check.
+
 2016-01-22  Nathan Sidwell  <nathan@codesourcery.com>
 
 	* omp-low.c (struct oacc_loop): Add 'inner' field.
Index: gcc/c/ChangeLog.gomp
===================================================================
--- gcc/c/ChangeLog.gomp	(revision 233104)
+++ gcc/c/ChangeLog.gomp	(working copy)
@@ -1,3 +1,7 @@
+2016-02-03  James Norris  <jnorris@codesourcery.com>
+
+	* c-typeck.c (build_external_ref): Remove usage check.
+
 2016-01-14  James Norris  <jnorris@codesourcery.com>
 
 	* c-parser.c (c_finish_oacc_routine): Remove attribute.
Index: gcc/c/c-typeck.c
===================================================================
--- gcc/c/c-typeck.c	(revision 233104)
+++ gcc/c/c-typeck.c	(working copy)
@@ -2677,26 +2677,6 @@
   tree ref;
   tree decl = lookup_name (id);
 
-  if (decl
-      && decl != error_mark_node
-      && current_function_decl
-      && TREE_CODE (decl) == VAR_DECL
-      && is_global_var (decl)
-      && get_oacc_fn_attrib (current_function_decl))
-    {
-      /* Validate data type for use with routine directive.  */
-      if (lookup_attribute ("omp declare target link",
-			    DECL_ATTRIBUTES (decl))
-	  || ((!lookup_attribute ("omp declare target",
-				  DECL_ATTRIBUTES (decl))
-	       && ((TREE_STATIC (decl) && !DECL_EXTERNAL (decl))
-		   || (!TREE_STATIC (decl) && DECL_EXTERNAL (decl))))))
-	{
-	  error_at (loc, "invalid use in %<routine%> function");
-	  return error_mark_node;
-	}
-    }
-
   /* In Objective-C, an instance variable (ivar) may be preferred to
      whatever lookup_name() found.  */
   decl = objc_lookup_ivar (decl, id);
Index: gcc/cp/ChangeLog.gomp
===================================================================
--- gcc/cp/ChangeLog.gomp	(revision 233104)
+++ gcc/cp/ChangeLog.gomp	(working copy)
@@ -1,3 +1,7 @@
+2016-02-03  James Norris  <jnorris@codesourcery.com>
+
+	* semantics.c (finish_id_expression): Remove usage check.
+
 2016-01-20  Cesar Philippidis  <cesar@codesourcery.com>
 
 	* parser.c (cp_parser_oacc_all_clauses): Call finish_omp_clauses
Index: gcc/cp/semantics.c
===================================================================
--- gcc/cp/semantics.c	(revision 233104)
+++ gcc/cp/semantics.c	(working copy)
@@ -3712,25 +3712,6 @@
 
 	  decl = convert_from_reference (decl);
 	}
-
-      if (decl != error_mark_node
-	  && current_function_decl
-	  && TREE_CODE (decl) == VAR_DECL
-	  && is_global_var (decl)
-          && get_oacc_fn_attrib (current_function_decl))
-	{
-	  /* Validate data type for use with routine directive.  */
-	  if (lookup_attribute ("omp declare target link",
-				DECL_ATTRIBUTES (decl))
-	      || ((!lookup_attribute ("omp declare target",
-				  DECL_ATTRIBUTES (decl))
-		   && ((TREE_STATIC (decl) && !DECL_EXTERNAL (decl))
-			|| (!TREE_STATIC (decl) && DECL_EXTERNAL (decl))))))
-	    {
-	      *error_msg = "invalid use in %<routine%> function";
-	      return error_mark_node;
-	    }
-	}
     }
 
   return cp_expr (decl, location);
Index: gcc/gimplify.c
===================================================================
--- gcc/gimplify.c	(revision 233104)
+++ gcc/gimplify.c	(working copy)
@@ -6095,9 +6095,9 @@
   if (ctx->region_type == ORT_NONE)
     return lang_hooks.decls.omp_disregard_value_expr (decl, false);
 
-  /* Threadprivate variables are predetermined.  */
   if (is_global_var (decl))
     {
+      /* Threadprivate variables are predetermined.  */
       if (DECL_THREAD_LOCAL_P (decl))
 	return omp_notice_threadprivate_variable (ctx, decl, NULL_TREE);
 
@@ -6108,6 +6108,30 @@
 	  if (value && DECL_P (value) && DECL_THREAD_LOCAL_P (value))
 	    return omp_notice_threadprivate_variable (ctx, decl, value);
 	}
+
+      if (gimplify_omp_ctxp->outer_context == NULL
+	  && VAR_P (decl)
+	  && get_oacc_fn_attrib (current_function_decl))
+	{
+	  location_t loc = DECL_SOURCE_LOCATION (decl);
+
+	  if (lookup_attribute ("omp declare target link",
+				DECL_ATTRIBUTES (decl)))
+	    {
+	      error_at (loc,
+			"%qE with %<link%> clause used in %<routine%> function",
+			DECL_NAME (decl));
+	      return false;
+	    }
+	  else if (!lookup_attribute ("omp declare target",
+				      DECL_ATTRIBUTES (decl)))
+	    {
+	      error_at (loc,
+			"%qE requires a %<declare%> directive for use "
+			"in a %<routine%> function", DECL_NAME (decl));
+	      return false;
+	    }
+	}
     }
 
   n = splay_tree_lookup (ctx->variables, (splay_tree_key)decl);
Index: gcc/testsuite/ChangeLog.gomp
===================================================================
--- gcc/testsuite/ChangeLog.gomp	(revision 233104)
+++ gcc/testsuite/ChangeLog.gomp	(working copy)
@@ -1,3 +1,9 @@
+2016-02-03  James Norris  <jnorris@codesourcery.com>
+
+	Backport from trunk:
+	2016-02-02  James Norris  <jnorris@codesourcery.com>
+	* c-c++-common/goacc/routine-5.c: Add tests.
+
 2016-01-22  Nathan Sidwell  <nathan@codesourcery.com>
 
 	* c-c++-common/goacc/loop-auto-1.c: Adjust expected warnings.
Index: gcc/testsuite/c-c++-common/goacc/routine-5.c
===================================================================
--- gcc/testsuite/c-c++-common/goacc/routine-5.c	(revision 233104)
+++ gcc/testsuite/c-c++-common/goacc/routine-5.c	(working copy)
@@ -60,48 +60,96 @@
 
 #pragma acc routine (Baz) // { dg-error "not been declared" }
 
-float vb1;
+int vb1;		/* { dg-error "directive for use" } */
+extern int vb2;		/* { dg-error "directive for use" } */
+static int vb3;		/* { dg-error "directive for use" } */
 
 #pragma acc routine
 int
 func1 (int a)
 {
-  vb1 = a + 1; /* { dg-error "invalid use in" } */
+  vb1 = a + 1;
+  vb2 = vb1 + 1;
+  vb3 = vb2 + 1;
 
-  return vb1; /* { dg-error "invalid use in" } */
+  return vb3;
 }
 
 #pragma acc routine
 int
 func2 (int a)
 {
-  static int vb2;
+  extern int vb4;	/* { dg-error "directive for use" } */
+  static int vb5;	/* { dg-error "directive for use" } */
 
-  vb2 = a + 1; /* { dg-error "invalid use in" } */
+  vb4 = a + 1;
+  vb5 = vb4 + 1;
 
-  return vb2; /* { dg-error "invalid use in" } */
+  return vb5;
 }
 
-float vb3;
-#pragma acc declare link (vb3)
+extern int vb6;			/* { dg-error "clause used in" } */
+#pragma acc declare link (vb6)
+static int vb7;			/* { dg-error "clause used in" } */
+#pragma acc declare link (vb7)
 
 #pragma acc routine
 int
 func3 (int a)
 {
-  vb3 = a + 1; /* { dg-error "invalid use in" } */
+  vb6 = a + 1;
+  vb7 = vb6 + 1;
 
-  return vb3; /* { dg-error "invalid use in" } */
+  return vb7;
 }
 
-float vb4;
-#pragma acc declare create (vb4)
+int vb8;
+#pragma acc declare create (vb8)
+extern int vb9;
+#pragma acc declare create (vb9)
+static int vb10;
+#pragma acc declare create (vb10)
 
 #pragma acc routine
 int
 func4 (int a)
 {
-  vb4 = a + 1;
+  vb8 = a + 1;
+  vb9 = vb8 + 1;
+  vb10 = vb9 + 1;
 
-  return vb4;
+  return vb10;
 }
+
+int vb11;
+#pragma acc declare device_resident (vb11)
+extern int vb12;
+#pragma acc declare device_resident (vb12)
+extern int vb13;
+#pragma acc declare device_resident (vb13)
+
+#pragma acc routine
+int
+func5 (int a)
+{
+  vb11 = a + 1;
+  vb12 = vb11 + 1;
+  vb13 = vb12 + 1;
+
+  return vb13;
+}
+
+#pragma acc routine
+int
+func6 (int a)
+{
+  extern int vb14;
+#pragma acc declare create (vb14)
+  static int vb15;
+#pragma acc declare create (vb15)
+
+  vb14 = a + 1;
+  vb15 = vb14 + 1;
+
+  return vb15;
+}