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]

[gomp4, committed] Revert "Add IFN_GOACC_DATA_END_WITH_ARG"


Hi,

I've reverted the patch that added IFN_GOACC_DATA_END_WITH_ARG ( https://gcc.gnu.org/ml/gcc-patches/2015-05/msg02661.html ).

The patch attempted to fix a test failure, while at the same time keeping the GOACC_data_start fnspec attributes to prevent it from becoming an alias analysis optimization barrier.

Now that we've got -foffload-alias, we're no longer concerned about GOACC builtins being alias analysis optimization barriers, so the IFN_GOACC_DATA_END_WITH_ARG patch has become obsolete.

Committed to gomp-4_0-branch.

Thanks,
- Tom
Revert "Add IFN_GOACC_DATA_END_WITH_ARG"

2015-10-05  Tom de Vries  <tom@codesourcery.com>

	revert:
	2015-05-28  Tom de Vries  <tom@codesourcery.com>

	PR tree-optimization/65419
	* cfgexpand.c (pass_data_expand): Add PROP_gimple_lompifn to
	properties_required field.
	* gimplify.c (gimplify_omp_workshare): Use IFN_GOACC_DATA_END_WITH_ARG
	instead of BUILT_IN_GOACC_DATA_END.  Clear PROP_gimple_lompifn in
	curr_properties.
	(gimplify_function_tree): Tentatively set PROP_gimple_lompifn in
	curr_properties.
	* internal-fn.c (expand_GOACC_DATA_END_WITH_ARG): New dummy function.
	* internal-fn.def (GOACC_DATA_END_WITH_ARG): New DEF_INTERNAL_FN.
	* omp-low.c (lower_omp_target): Set argument of GOACC_DATA_END_WITH_ARG.
	(pass_data_late_lower_omp): New pass_data.
	(pass_late_lower_omp): New pass.
	(pass_late_lower_omp::gate, pass_late_lower_omp::execute)
	(make_pass_late_lower_omp): New function.
	* passes.def: Add pass_late_lower_omp.
	* tree-inline.c (expand_call_inline): Handle PROP_gimple_lompifn.
	* tree-pass.h (PROP_gimple_lompifn): Add define.

	* testsuite/libgomp.oacc-c-c++-common/goacc-data-end.c: New test.
---
 gcc/cfgexpand.c                                    |  3 +-
 gcc/gimplify.c                                     | 25 ++-----
 gcc/internal-fn.c                                  |  9 ---
 gcc/internal-fn.def                                |  1 -
 gcc/omp-low.c                                      | 86 +---------------------
 gcc/passes.def                                     |  1 -
 gcc/tree-inline.c                                  | 16 ++--
 gcc/tree-pass.h                                    |  2 -
 .../libgomp.oacc-c-c++-common/goacc-data-end.c     | 67 -----------------
 9 files changed, 14 insertions(+), 196 deletions(-)
 delete mode 100644 libgomp/testsuite/libgomp.oacc-c-c++-common/goacc-data-end.c

diff --git a/gcc/cfgexpand.c b/gcc/cfgexpand.c
index ca52d3d..bfbc958 100644
--- a/gcc/cfgexpand.c
+++ b/gcc/cfgexpand.c
@@ -6060,8 +6060,7 @@ const pass_data pass_data_expand =
   ( PROP_ssa | PROP_gimple_leh | PROP_cfg
     | PROP_gimple_lcx
     | PROP_gimple_lvec
-    | PROP_gimple_lva
-    | PROP_gimple_lompifn), /* properties_required */
+    | PROP_gimple_lva), /* properties_required */
   PROP_rtl, /* properties_provided */
   ( PROP_ssa | PROP_trees ), /* properties_destroyed */
   0, /* todo_flags_start */
diff --git a/gcc/gimplify.c b/gcc/gimplify.c
index 6283f0c..a5e28b4 100644
--- a/gcc/gimplify.c
+++ b/gcc/gimplify.c
@@ -8960,32 +8960,20 @@ gimplify_omp_workshare (tree *expr_p, gimple_seq *pre_p)
 	pop_gimplify_context (NULL);
       if (ort == ORT_TARGET_DATA)
 	{
+	  enum built_in_function end_ix;
 	  switch (TREE_CODE (expr))
 	    {
 	    case OACC_DATA:
-	      /* Rather than building a call to BUILT_IN_GOACC_DATA_END, we use
-		 this ifn which is similar, but has a pointer argument, which
-		 will be later set to the &.omp_data_arr of the corresponding
-		 BUILT_IN_GOACC_DATA_START.
-		 This allows us to pretend that the &.omp_data_arr argument of
-		 BUILT_IN_GOACC_DATA_START does not escape.  */
-	      g = gimple_build_call_internal (IFN_GOACC_DATA_END_WITH_ARG, 1,
-					      null_pointer_node);
-	      /* Clear the tentatively set PROP_gimple_lompifn, to indicate that
-		 IFN_GOACC_DATA_END_WITH_ARG needs to be expanded.  The argument
-		 is not abi-compatible with the GOACC_data_end function, which
-		 has no arguments.  */
-	      cfun->curr_properties &= ~PROP_gimple_lompifn;
+	      end_ix = BUILT_IN_GOACC_DATA_END;
 	      break;
 	    case OMP_TARGET_DATA:
-	      {
-		tree fn = builtin_decl_explicit (BUILT_IN_GOMP_TARGET_END_DATA);
-		g = gimple_build_call (fn, 0);
-	      }
+	      end_ix = BUILT_IN_GOMP_TARGET_END_DATA;
 	      break;
 	    default:
 	      gcc_unreachable ();
 	    }
+	  tree fn = builtin_decl_explicit (end_ix);
+	  g = gimple_build_call (fn, 0);
 	  gimple_seq cleanup = NULL;
 	  gimple_seq_add_stmt (&cleanup, g);
 	  g = gimple_build_try (body, cleanup, GIMPLE_TRY_FINALLY);
@@ -10939,9 +10927,6 @@ gimplify_function_tree (tree fndecl)
      if necessary.  */
   cfun->curr_properties |= PROP_gimple_lva;
 
-  /* Tentatively set PROP_gimple_lompifn.  */
-  cfun->curr_properties |= PROP_gimple_lompifn;
-
   for (parm = DECL_ARGUMENTS (fndecl); parm ; parm = DECL_CHAIN (parm))
     {
       /* Preliminarily mark non-addressed complex variables as eligible
diff --git a/gcc/internal-fn.c b/gcc/internal-fn.c
index 317149e..6fac752 100644
--- a/gcc/internal-fn.c
+++ b/gcc/internal-fn.c
@@ -1949,15 +1949,6 @@ expand_VA_ARG (gcall *stmt ATTRIBUTE_UNUSED)
   gcc_unreachable ();
 }
 
-/* GOACC_DATA_END_WITH_ARG is supposed to be expanded at pass_late_lower_omp.
-   So this dummy function should never be called.  */
-
-static void
-expand_GOACC_DATA_END_WITH_ARG (gcall *stmt ATTRIBUTE_UNUSED)
-{
-  gcc_unreachable ();
-}
-
 /* Expand the IFN_UNIQUE function according to its first argument.  */
 
 static void
diff --git a/gcc/internal-fn.def b/gcc/internal-fn.def
index 524b98d..ca06b10 100644
--- a/gcc/internal-fn.def
+++ b/gcc/internal-fn.def
@@ -65,7 +65,6 @@ DEF_INTERNAL_FN (SUB_OVERFLOW, ECF_CONST | ECF_LEAF | ECF_NOTHROW, NULL)
 DEF_INTERNAL_FN (MUL_OVERFLOW, ECF_CONST | ECF_LEAF | ECF_NOTHROW, NULL)
 DEF_INTERNAL_FN (TSAN_FUNC_EXIT, ECF_NOVOPS | ECF_LEAF | ECF_NOTHROW, NULL)
 DEF_INTERNAL_FN (VA_ARG, ECF_NOTHROW | ECF_LEAF, NULL)
-DEF_INTERNAL_FN (GOACC_DATA_END_WITH_ARG, ECF_NOTHROW, ".r")
 
 /* An unduplicable, uncombinable function.  Generally used to preserve
    a CFG property in the face of jump threading, tail merging or
diff --git a/gcc/omp-low.c b/gcc/omp-low.c
index debedb1..130728a 100644
--- a/gcc/omp-low.c
+++ b/gcc/omp-low.c
@@ -14999,7 +14999,6 @@ lower_omp_target (gimple_stmt_iterator *gsi_p, omp_context *ctx)
   bool offloaded, data_region;
   unsigned int map_cnt = 0;
   bool has_depend = false;
-  gimple *goacc_data_end = NULL;
 
   offloaded = is_gimple_omp_offloaded (stmt);
   switch (gimple_omp_target_kind (stmt))
@@ -15044,18 +15043,7 @@ lower_omp_target (gimple_stmt_iterator *gsi_p, omp_context *ctx)
       tgt_body = gimple_bind_body (tgt_bind);
     }
   else if (data_region)
-    {
-      tgt_body = gimple_omp_body (stmt);
-      gimple *try_stmt = gimple_seq_first_stmt (tgt_body);
-      gcc_assert (gimple_try_kind (try_stmt) == GIMPLE_TRY_FINALLY);
-      gimple_seq cleanup = gimple_try_cleanup (try_stmt);
-      if (gimple_call_internal_p (cleanup)
-	  && gimple_call_internal_fn (cleanup) == IFN_GOACC_DATA_END_WITH_ARG)
-	{
-	  goacc_data_end = cleanup;
-	  gcc_assert (gimple_call_arg (goacc_data_end, 0) == null_pointer_node);
-	}
-    }
+    tgt_body = gimple_omp_body (stmt);
   child_fn = ctx->cb.dst_fn;
 
   push_gimplify_context ();
@@ -15296,13 +15284,6 @@ lower_omp_target (gimple_stmt_iterator *gsi_p, omp_context *ctx)
 	= create_tmp_var (ctx->record_type, ".omp_data_arr");
       DECL_NAMELESS (ctx->sender_decl) = 1;
       TREE_ADDRESSABLE (ctx->sender_decl) = 1;
-
-      if (goacc_data_end != NULL)
-	{
-	  tree arg = build_fold_addr_expr (ctx->sender_decl);
-	  gimple_call_set_arg (goacc_data_end, 0, arg);
-	}
-
       t = make_tree_vec (3);
       TREE_VEC_ELT (t, 0) = ctx->sender_decl;
       TREE_VEC_ELT (t, 1)
@@ -18547,71 +18528,6 @@ omp_finish_file (void)
     }
 }
 
-namespace {
-
-const pass_data pass_data_late_lower_omp =
-{
-  GIMPLE_PASS, /* type */
-  "lateomplower", /* name */
-  OPTGROUP_NONE, /* optinfo_flags */
-  TV_NONE, /* tv_id */
-  ( PROP_cfg | PROP_ssa ), /* properties_required */
-  PROP_gimple_lompifn, /* properties_provided */
-  0, /* properties_destroyed */
-  0, /* todo_flags_start */
-  0, /* todo_flags_finish */
-};
-
-class pass_late_lower_omp : public gimple_opt_pass
-{
-public:
-  pass_late_lower_omp (gcc::context *ctxt)
-    : gimple_opt_pass (pass_data_late_lower_omp, ctxt)
-  {}
-
-  /* opt_pass methods: */
-  virtual unsigned int execute (function *);
-
-  virtual bool gate (function *)
-    {
-      return (cfun->curr_properties & PROP_gimple_lompifn) == 0;
-    }
-
-}; // class pass_lower_omp
-
-unsigned int
-pass_late_lower_omp::execute (function *fun)
-{
-  basic_block bb;
-  gimple_stmt_iterator i;
-
-  FOR_EACH_BB_FN (bb, fun)
-    for (i = gsi_start_bb (bb); !gsi_end_p (i); gsi_next (&i))
-      {
-	gimple *stmt = gsi_stmt (i);
-	if (!(is_gimple_call (stmt)
-	      && gimple_call_internal_p (stmt)
-	      && gimple_call_internal_fn (stmt) == IFN_GOACC_DATA_END_WITH_ARG))
-	  continue;
-
-	tree fn = builtin_decl_explicit (BUILT_IN_GOACC_DATA_END);
-	gimple *g = gimple_build_call (fn, 0);
-
-	gsi_replace (&i, g, false);
-      }
-
-  return TODO_update_ssa;
-}
-
-
-} // anon namespace
-
-gimple_opt_pass *
-make_pass_late_lower_omp (gcc::context *ctxt)
-{
-  return new pass_late_lower_omp (ctxt);
-}
-
 /* Transform oacc_dim_size and oacc_dim_pos internal function calls to
    constants, where possible.  */
 
diff --git a/gcc/passes.def b/gcc/passes.def
index e68228e..5683bb7 100644
--- a/gcc/passes.def
+++ b/gcc/passes.def
@@ -367,7 +367,6 @@ along with GCC; see the file COPYING3.  If not see
   POP_INSERT_PASSES ()
   NEXT_PASS (pass_simduid_cleanup);
   NEXT_PASS (pass_vtable_verify);
-  NEXT_PASS (pass_late_lower_omp);
   NEXT_PASS (pass_lower_vaarg);
   NEXT_PASS (pass_lower_vector);
   NEXT_PASS (pass_lower_complex_O0);
diff --git a/gcc/tree-inline.c b/gcc/tree-inline.c
index 00ad2af..3d06e6e 100644
--- a/gcc/tree-inline.c
+++ b/gcc/tree-inline.c
@@ -4540,15 +4540,13 @@ expand_call_inline (basic_block bb, gimple *stmt, copy_body_data *id)
   id->src_cfun = DECL_STRUCT_FUNCTION (fn);
   id->call_stmt = stmt;
 
-  {
-    /* Handle properties that need to be false in the resulting function, if
-       they're false in the src function.  */
-    unsigned int props_mask = PROP_gimple_lva | PROP_gimple_lompifn;
-    unsigned int src_props = id->src_cfun->curr_properties;
-    unsigned int kill_props = props_mask & ~src_props;
-    struct function *dst_cfun = DECL_STRUCT_FUNCTION (id->dst_fn);
-    dst_cfun->curr_properties &= ~kill_props;
-  }
+  /* If the the src function contains an IFN_VA_ARG, then so will the dst
+     function after inlining.  */
+  if ((id->src_cfun->curr_properties & PROP_gimple_lva) == 0)
+    {
+      struct function *dst_cfun = DECL_STRUCT_FUNCTION (id->dst_fn);
+      dst_cfun->curr_properties &= ~PROP_gimple_lva;
+    }
 
   gcc_assert (!id->src_cfun->after_inlining);
 
diff --git a/gcc/tree-pass.h b/gcc/tree-pass.h
index 248c406..8eaf678 100644
--- a/gcc/tree-pass.h
+++ b/gcc/tree-pass.h
@@ -226,7 +226,6 @@ protected:
 						   of math functions; the
 						   current choices have
 						   been optimized.  */
-#define PROP_gimple_lompifn	(1 << 16)       /* No omp internal function.  */
 
 #define PROP_trees \
   (PROP_gimple_any | PROP_gimple_lcf | PROP_gimple_leh | PROP_gimple_lomp)
@@ -415,7 +414,6 @@ extern gimple_opt_pass *make_pass_lower_complex (gcc::context *ctxt);
 extern gimple_opt_pass *make_pass_lower_vector (gcc::context *ctxt);
 extern gimple_opt_pass *make_pass_lower_vector_ssa (gcc::context *ctxt);
 extern gimple_opt_pass *make_pass_lower_omp (gcc::context *ctxt);
-extern gimple_opt_pass *make_pass_late_lower_omp (gcc::context *ctxt);
 extern gimple_opt_pass *make_pass_diagnose_omp_blocks (gcc::context *ctxt);
 extern gimple_opt_pass *make_pass_expand_omp (gcc::context *ctxt);
 extern gimple_opt_pass *make_pass_expand_omp_ssa (gcc::context *ctxt);
diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/goacc-data-end.c b/libgomp/testsuite/libgomp.oacc-c-c++-common/goacc-data-end.c
deleted file mode 100644
index 5685dbd..0000000
--- a/libgomp/testsuite/libgomp.oacc-c-c++-common/goacc-data-end.c
+++ /dev/null
@@ -1,67 +0,0 @@
-/* { dg-do run } */
-
-/* Data directive at end of function.  This is the variant that triggered
-   PR65419.  */
-
-static void __attribute__((noinline,noclone))
-f (void)
-{
-  int i;
-
-#pragma acc data copyout (i)
-  {
-
-  }
-}
-
-/* Data directive in inlined function g_1.  */
-
-static inline void
-g_1 (void)
-{
-  int i;
-
-#pragma acc data copyout (i)
-  {
-
-  }
-}
-
-static void __attribute__((noinline,noclone))
-g (void)
-{
-  g_1 ();
-}
-
-/* Data directive in function h into which a function h_1 is inlined.  */
-
-static inline void
-h_1 (void)
-{
-
-}
-
-static void __attribute__((noinline,noclone))
-h (void)
-{
-  int i;
-
-  h_1 ();
-
-#pragma acc data copyout (i)
-  {
-
-  }
-}
-
-/* Main function calling the tests.  */
-
-int
-main (void)
-{
-  f ();
-  g ();
-  h ();
-
-  return 0;
-}
-- 
1.9.1


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