This is the mail archive of the gcc-patches@gcc.gnu.org mailing list for the GCC project.


Index Nav: [Date Index] [Subject Index] [Author Index] [Thread Index]
Message Nav: [Date Prev] [Date Next] [Thread Prev] [Thread Next]
Other format: [Raw text]

Re: [PATCH] Fix use of declare'd vars by routine procedures.


Jakub,

On 01/22/2016 04:39 AM, Jakub Jelinek wrote:

As for the extra error, I think it would be better to diagnose this during
gimplification, that way you do it for all FEs, plus you have the omp
context in there, etc.  The error wording is weird, the diagnostic
should make it clear use of what is invalid in what and why.

	Jakub


I think the attached change is what you had in mind with
regard to doing the check at gimplification time.

OK?

Thanks for taking the time to review,
Jim


ChangeLog entries....

        gcc/
        * gimplify.c (gimplify_var_or_parm_decl): Add variable usage check.
        gcc/testsuite/
        * c-c++-common/goacc/routine-5.c: Add tests.

Index: gcc/ChangeLog
===================================================================
diff --git a/trunk/gcc/ChangeLog b/trunk/gcc/ChangeLog
--- a/trunk/gcc/ChangeLog	(revision 232802)
+++ b/trunk/gcc/ChangeLog	(working copy)
@@ -1,3 +1,7 @@
+2016-01-XX  James Norris  <jnorris@codesourcery.com>
+
+	* gimplify.c (gimplify_var_or_parm_decl): Add usage check.
+
 2016-01-25  Jeff Law  <law@redhat.com>
 
 	PR tree-optimization/69196
Index: gcc/gimplify.c
===================================================================
diff --git a/trunk/gcc/gimplify.c b/trunk/gcc/gimplify.c
--- a/trunk/gcc/gimplify.c	(revision 232802)
+++ b/trunk/gcc/gimplify.c	(working copy)
@@ -1841,6 +1841,33 @@
       return GS_ERROR;
     }
 
+  /* Validate variable for use within routine function.  */
+  if (gimplify_omp_ctxp && gimplify_omp_ctxp->region_type == ORT_TARGET
+      && get_oacc_fn_attrib (current_function_decl)
+      && TREE_CODE (decl) == VAR_DECL
+      && is_global_var (decl)
+      && ((TREE_STATIC (decl) && !DECL_EXTERNAL (decl))
+	  || (!TREE_STATIC (decl) && DECL_EXTERNAL (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 GS_ERROR;
+	}
+      else if (!lookup_attribute ("omp declare target", DECL_ATTRIBUTES (decl)))
+	{
+	  error_at (loc,
+		    "storage class of %qE cannot be ", DECL_NAME (decl));
+	  error_at (gimplify_omp_ctxp->location,
+		    "used in enclosing %<routine%> function");
+	  return GS_ERROR;
+	}
+    }
+
   /* When within an OMP context, notice uses of variables.  */
   if (gimplify_omp_ctxp && omp_notice_variable (gimplify_omp_ctxp, decl, true))
     return GS_ALL_DONE;
Index: gcc/testsuite/ChangeLog
===================================================================
diff --git a/trunk/gcc/testsuite/ChangeLog b/trunk/gcc/testsuite/ChangeLog
--- a/trunk/gcc/testsuite/ChangeLog	(revision 232802)
+++ b/trunk/gcc/testsuite/ChangeLog	(working copy)
@@ -1,3 +1,7 @@
+2016-01-XX  James Norris  <jnorris@codesourcery.com>
+
+	* c-c++-common/goacc/routine-5.c: Add tests.
+
 2016-01-25  Jeff Law  <law@redhat.com>
 
 	PR tree-optimization/69196
Index: gcc/testsuite/c-c++-common/goacc/routine-5.c
===================================================================
diff --git a/trunk/gcc/testsuite/c-c++-common/goacc/routine-5.c b/trunk/gcc/testsuite/c-c++-common/goacc/routine-5.c
--- a/trunk/gcc/testsuite/c-c++-common/goacc/routine-5.c	(revision 232802)
+++ b/trunk/gcc/testsuite/c-c++-common/goacc/routine-5.c	(working copy)
@@ -45,3 +45,70 @@
 #pragma acc routine (a) /* { dg-error "does not refer to" } */
   
 #pragma acc routine (c) /* { dg-error "does not refer to" } */
+
+float vb1; /* { dg-error "storage class of" } */
+
+#pragma acc routine
+int
+func1 (int a)		/* { dg-error "used in enclosing" } */
+{
+  vb1 = a + 1;
+
+  return vb1;
+}
+
+#pragma acc routine
+int
+func2 (int a)		/* { dg-error "used in enclosing" } */
+{
+  static int vb2;	/* { dg-error "storage class of" } */
+
+  vb2 = a + 1;
+
+  return vb2;
+}
+
+float vb3;		/* { dg-error "'link' clause used" } */
+#pragma acc declare link (vb3)
+
+#pragma acc routine
+int
+func3 (int a)
+{
+  vb3 = a + 1;
+
+  return vb3;
+}
+
+float vb4;
+#pragma acc declare create (vb4)
+
+#pragma acc routine
+int
+func4 (int a)
+{
+  vb4 = a + 1;
+
+  return vb4;
+}
+
+extern float vb5;	/* { dg-error "storage class of" } */
+
+#pragma acc routine
+int
+func5 (int a)		/* { dg-error "used in enclosing" } */
+{
+  vb5 = a + 1;
+
+  return vb5;
+}
+
+#pragma acc routine
+int
+func6 (int a)		/* { dg-error "used in enclosing" } */
+{
+  extern float vb6;	/* { dg-error "storage class of" } */
+  vb6 = a + 1;
+
+  return vb6;
+}

Index Nav: [Date Index] [Subject Index] [Author Index] [Thread Index]
Message Nav: [Date Prev] [Date Next] [Thread Prev] [Thread Next]