maybe_fold_stmt (was: [gomp4] #pragma omp target* fixes)

[Thomas Schwinge <thomas at codesourcery dot com>]

Hi!

On Thu, 5 Sep 2013 18:11:05 +0200, Jakub Jelinek <jakub@redhat.com> wrote:
> 4) the reference testcase showed a problem with fold_stmt calls
> we do very early, during gimplification, because for TREE_READONLY
> vars with DECL_INITIAL fold_stmt can replace the uses of the var with
> its initializer, but as the gimplifier isn't aware of it, we wouldn't remap
> that, or worse there could be explicit remapping of it via array section,
> but one that the compiler doesn't see, and if that is smaller than
> the whole array size, that would result in runtime error.  So, after
> some talk with richi on IRC, I've decided to just not fold_stmt
> inside of target constructs during gimplification and defer it until
> omplower.

> 	* gimplify.c (gimplify_call_expr): Don't call fold_stmt
> 	inside of #pragma omp target construct.
> 	(gimplify_modify_expr): Likewise.
> 	* omp-low.c
> 	(lower_omp): Call fold_stmt on all stmts inside of
> 	#pragma omp target construct.

> --- gcc/gimplify.c.jj	2013-09-05 09:19:03.000000000 +0200
> +++ gcc/gimplify.c	2013-09-05 14:45:48.632720617 +0200
> @@ -2704,7 +2704,14 @@ gimplify_call_expr (tree *expr_p, gimple
>        notice_special_calls (call);
>        gimplify_seq_add_stmt (pre_p, call);
>        gsi = gsi_last (*pre_p);
> -      fold_stmt (&gsi);
> +      /* Don't fold stmts inside of target construct.  We'll do it
> +	 during omplower pass instead.  */
> +      struct gimplify_omp_ctx *ctx;
> +      for (ctx = gimplify_omp_ctxp; ctx; ctx = ctx->outer_context)
> +	if (ctx->region_type == ORT_TARGET)
> +	  break;
> +      if (ctx == NULL)
> +	fold_stmt (&gsi);
>        *expr_p = NULL_TREE;
>      }
>    else
> @@ -4961,7 +4968,14 @@ gimplify_modify_expr (tree *expr_p, gimp
>  
>    gimplify_seq_add_stmt (pre_p, assign);
>    gsi = gsi_last (*pre_p);
> -  fold_stmt (&gsi);
> +  /* Don't fold stmts inside of target construct.  We'll do it
> +     during omplower pass instead.  */
> +  struct gimplify_omp_ctx *ctx;
> +  for (ctx = gimplify_omp_ctxp; ctx; ctx = ctx->outer_context)
> +    if (ctx->region_type == ORT_TARGET)
> +      break;
> +  if (ctx == NULL)
> +    fold_stmt (&gsi);
>  
>    if (want_value)
>      {
> --- gcc/omp-low.c.jj	2013-09-05 09:19:03.000000000 +0200
> +++ gcc/omp-low.c	2013-09-05 17:11:14.693638660 +0200
> @@ -9673,6 +9685,12 @@ lower_omp (gimple_seq *body, omp_context
>    gimple_stmt_iterator gsi;
>    for (gsi = gsi_start (*body); !gsi_end_p (gsi); gsi_next (&gsi))
>      lower_omp_1 (&gsi, ctx);
> +  /* Inside target region we haven't called fold_stmt during gimplification,
> +     because it can break code by adding decl references that weren't in the
> +     source.  Call fold_stmt now.  */
> +  if (target_nesting_level)
> +    for (gsi = gsi_start (*body); !gsi_end_p (gsi); gsi_next (&gsi))
> +      fold_stmt (&gsi);
>    input_location = saved_location;
>  }

OK to commit to trunk the following patch?

commit 6ff10eb51ea39a25e53e0369626559be208bb16e
Author: Thomas Schwinge <thomas@codesourcery.com>
Date:   Mon Oct 21 16:37:41 2013 +0200

    Refactor common code into new maybe_fold_stmt function.
    
    	gcc/
    	* gimplify.c (gimplify_call_expr, gimplify_modify_expr): Move
    	common code...
    	(maybe_fold_stmt): ... into this new function.
    	* omp-low.c (lower_omp): Update comment.

diff --git gcc/gimplify.c gcc/gimplify.c
index 1ca847a..7203456 100644
--- gcc/gimplify.c
+++ gcc/gimplify.c
@@ -2184,6 +2184,23 @@ gimplify_arg (tree *arg_p, gimple_seq *pre_p, location_t call_location)
   return gimplify_expr (arg_p, pre_p, NULL, test, fb);
 }
 
+/* Don't fold STMT inside ORT_TARGET, because it can break code by adding decl
+   references that weren't in the source.  We'll do it during omplower pass
+   instead.  */
+
+static bool
+maybe_fold_stmt (gimple_stmt_iterator *gsi)
+{
+  bool changed = false;
+  struct gimplify_omp_ctx *ctx;
+  for (ctx = gimplify_omp_ctxp; ctx; ctx = ctx->outer_context)
+    if (ctx->region_type == ORT_TARGET)
+      break;
+  if (ctx == NULL)
+    changed = fold_stmt (gsi);
+  return changed;
+}
+
 /* Gimplify the CALL_EXPR node *EXPR_P into the GIMPLE sequence PRE_P.
    WANT_VALUE is true if the result of the call is desired.  */
 
@@ -2417,14 +2434,7 @@ gimplify_call_expr (tree *expr_p, gimple_seq *pre_p, bool want_value)
       notice_special_calls (call);
       gimplify_seq_add_stmt (pre_p, call);
       gsi = gsi_last (*pre_p);
-      /* Don't fold stmts inside of target construct.  We'll do it
-	 during omplower pass instead.  */
-      struct gimplify_omp_ctx *ctx;
-      for (ctx = gimplify_omp_ctxp; ctx; ctx = ctx->outer_context)
-	if (ctx->region_type == ORT_TARGET)
-	  break;
-      if (ctx == NULL)
-	fold_stmt (&gsi);
+      maybe_fold_stmt (&gsi);
       *expr_p = NULL_TREE;
     }
   else
@@ -4572,14 +4582,7 @@ gimplify_modify_expr (tree *expr_p, gimple_seq *pre_p, gimple_seq *post_p,
 
   gimplify_seq_add_stmt (pre_p, assign);
   gsi = gsi_last (*pre_p);
-  /* Don't fold stmts inside of target construct.  We'll do it
-     during omplower pass instead.  */
-  struct gimplify_omp_ctx *ctx;
-  for (ctx = gimplify_omp_ctxp; ctx; ctx = ctx->outer_context)
-    if (ctx->region_type == ORT_TARGET)
-      break;
-  if (ctx == NULL)
-    fold_stmt (&gsi);
+  maybe_fold_stmt (&gsi);
 
   if (want_value)
     {
diff --git gcc/omp-low.c gcc/omp-low.c
index 05fca40..1222df6 100644
--- gcc/omp-low.c
+++ gcc/omp-low.c
@@ -10110,9 +10110,8 @@ lower_omp (gimple_seq *body, omp_context *ctx)
   gimple_stmt_iterator gsi;
   for (gsi = gsi_start (*body); !gsi_end_p (gsi); gsi_next (&gsi))
     lower_omp_1 (&gsi, ctx);
-  /* Inside target region we haven't called fold_stmt during gimplification,
-     because it can break code by adding decl references that weren't in the
-     source.  Call fold_stmt now.  */
+  /* During gimplification, we haven't called fold_stmt inside ORT_TARGET
+     (gimplify.c:maybe_fold_stmt); call it now.  */
   if (target_nesting_level)
     for (gsi = gsi_start (*body); !gsi_end_p (gsi); gsi_next (&gsi))
       fold_stmt (&gsi);


GrÃÃe,
 Thomas

Attachment: pgpsr38OVdggl.pgp
Description: PGP signature