diff mbox

Untested #pragma acc declare fix

Message ID 20161007122132.GG7282@tucnak.redhat.com
State New
Headers show

Commit Message

Jakub Jelinek Oct. 7, 2016, 12:21 p.m. UTC 
Hi!

During review of Martin's -fsanitize-use-after-scope patch, I've noticed
what I believe is a bug in #pragma acc declare support.

In particular, the oacc_declare_returns has been added next to the CLOBBER
additions, but the clobbers are guarded with many conditions, e.g. aren't
emitted with -fstack-reuse=none and many other conditions.

I believe the additions to that data structure aren't guarded by these
conditions, it is just for VAR_DECLs that aren't is_global_var and have
current_function_decl context.

I haven't tested this patch (can bootstrap/regtest it), but don't have time
to try to write a testcase for all the cases where the conditions matter.

2016-10-07  Jakub Jelinek  <jakub@redhat.com>

	* gimplify.c (gimplify_bind_expr): Handle oacc_declare_returns
	even for -fstack-reuse=none, or for volatile vars etc.


	Jakub

Comments

Nathan Sidwell Oct. 7, 2016, 7:45 p.m. UTC | #1 
On 10/07/16 08:21, Jakub Jelinek wrote:
> Hi!
>
> During review of Martin's -fsanitize-use-after-scope patch, I've noticed
> what I believe is a bug in #pragma acc declare support.
>
> In particular, the oacc_declare_returns has been added next to the CLOBBER
> additions, but the clobbers are guarded with many conditions, e.g. aren't
> emitted with -fstack-reuse=none and many other conditions.
>
> I believe the additions to that data structure aren't guarded by these
> conditions, it is just for VAR_DECLs that aren't is_global_var and have
> current_function_decl context.
>
> I haven't tested this patch (can bootstrap/regtest it), but don't have time
> to try to write a testcase for all the cases where the conditions matter.

thanks, your logic seems sound.  I'll see what we can do about a test case.

nathan
diff mbox

Patch

--- gcc/gimplify.c.jj	2016-09-26 12:06:46.000000000 +0200
+++ gcc/gimplify.c	2016-10-07 14:11:09.299624104 +0200
@@ -1192,21 +1192,24 @@  gimplify_bind_expr (tree *expr_p, gimple
     {
       if (TREE_CODE (t) == VAR_DECL
 	  && !is_global_var (t)
-	  && DECL_CONTEXT (t) == current_function_decl
-	  && !DECL_HARD_REGISTER (t)
-	  && !TREE_THIS_VOLATILE (t)
-	  && !DECL_HAS_VALUE_EXPR_P (t)
-	  /* Only care for variables that have to be in memory.  Others
-	     will be rewritten into SSA names, hence moved to the top-level.  */
-	  && !is_gimple_reg (t)
-	  && flag_stack_reuse != SR_NONE)
+	  && DECL_CONTEXT (t) == current_function_decl)
 	{
-	  tree clobber = build_constructor (TREE_TYPE (t), NULL);
-	  gimple *clobber_stmt;
-	  TREE_THIS_VOLATILE (clobber) = 1;
-	  clobber_stmt = gimple_build_assign (t, clobber);
-	  gimple_set_location (clobber_stmt, end_locus);
-	  gimplify_seq_add_stmt (&cleanup, clobber_stmt);
+	  if (!DECL_HARD_REGISTER (t)
+	      && !TREE_THIS_VOLATILE (t)
+	      && !DECL_HAS_VALUE_EXPR_P (t)
+	      /* Only care for variables that have to be in memory.  Others
+		 will be rewritten into SSA names, hence moved to the
+		 top-level.  */
+	      && !is_gimple_reg (t)
+	      && flag_stack_reuse != SR_NONE)
+	    {
+	      tree clobber = build_constructor (TREE_TYPE (t), NULL);
+	      gimple *clobber_stmt;
+	      TREE_THIS_VOLATILE (clobber) = 1;
+	      clobber_stmt = gimple_build_assign (t, clobber);
+	      gimple_set_location (clobber_stmt, end_locus);
+	      gimplify_seq_add_stmt (&cleanup, clobber_stmt);
+	    }
 
 	  if (flag_openacc && oacc_declare_returns != NULL)
 	    {