diff mbox

[Bulk,OpenACC,0/7] host_data construct

Message ID 87twn9h1hg.fsf@kepler.schwinge.homeip.net
State New
Headers show

Commit Message

Thomas Schwinge Dec. 23, 2015, 11:02 a.m. UTC 
Hi!

On Mon, 26 Oct 2015 19:34:22 +0100, Jakub Jelinek <jakub@redhat.com> wrote:
> Your use_device sounds very similar to use_device_ptr clause in OpenMP,
> which is allowed on #pragma omp target data construct and is implemented
> quite a bit differently from this; it is unclear if the OpenACC standard
> requires this kind of implementation, or you just chose to implement it this
> way.  In particular, the GOMP_target_data call puts the variables mentioned
> in the use_device_ptr clauses into the mapping structures (similarly how
> map clause appears) and the corresponding vars are privatized within the
> target data region (which is a host region, basically a fancy { } braces),
> where the private variables contain the offloading device's pointers.

ACK.  As the OpenACC use_device clause implementation now completely
matches the OpenMP use_device_ptr clause implementation, there is no use
anymore for a separate OMP_CLAUSE_USE_DEVICE, so in r231926 I cleaned
that up, as obvious:

commit 9d5fd7c608fef6e7a9efbfc940545d49452c4e01
Author: tschwinge <tschwinge@138bc75d-0d04-0410-961f-82ee72b054a4>
Date:   Wed Dec 23 11:01:18 2015 +0000

    Merge OMP_CLAUSE_USE_DEVICE into OMP_CLAUSE_USE_DEVICE_PTR
    
    	gcc/c/
    	* c-parser.c (c_parser_oacc_clause_use_device): Merge function
    	into...
    	(c_parser_omp_clause_use_device_ptr): ... this function.  Adjust
    	all users.
    	gcc/
    	* tree-core.h (enum omp_clause_code): Merge OMP_CLAUSE_USE_DEVICE
    	into OMP_CLAUSE_USE_DEVICE_PTR.  Adjust all users.
    
    git-svn-id: svn+ssh://gcc.gnu.org/svn/gcc/trunk@231926 138bc75d-0d04-0410-961f-82ee72b054a4
---
 gcc/ChangeLog                                      |  5 +++++
 gcc/c/ChangeLog                                    |  7 +++++++
 gcc/c/c-parser.c                                   | 16 +++++-----------
 gcc/c/c-typeck.c                                   |  1 -
 gcc/cp/parser.c                                    |  2 +-
 gcc/cp/pt.c                                        |  2 --
 gcc/cp/semantics.c                                 |  1 -
 gcc/fortran/trans-openmp.c                         |  2 +-
 gcc/gimplify.c                                     |  2 --
 gcc/omp-low.c                                      | 11 ++---------
 gcc/testsuite/gfortran.dg/goacc/host_data-tree.f95 |  2 +-
 gcc/tree-core.h                                    |  6 ++----
 gcc/tree-nested.c                                  |  2 --
 gcc/tree-pretty-print.c                            |  3 ---
 gcc/tree.c                                         |  3 ---
 15 files changed, 24 insertions(+), 41 deletions(-)



Grüße
 Thomas
diff mbox

Patch

diff --git gcc/ChangeLog gcc/ChangeLog
index aa28d10..d67b9c6 100644
--- gcc/ChangeLog
+++ gcc/ChangeLog
@@ -1,3 +1,8 @@ 
+2015-12-23  Thomas Schwinge  <thomas@codesourcery.com>
+
+	* tree-core.h (enum omp_clause_code): Merge OMP_CLAUSE_USE_DEVICE
+	into OMP_CLAUSE_USE_DEVICE_PTR.  Adjust all users.
+
 2015-12-23  David Sherwood  <david.sherwood@arm.com>
 
 	* config/arm/iterators.md (VMAXMINFNM): New int iterator.
diff --git gcc/c/ChangeLog gcc/c/ChangeLog
index f99f426..7b275d8 100644
--- gcc/c/ChangeLog
+++ gcc/c/ChangeLog
@@ -1,3 +1,10 @@ 
+2015-12-23  Thomas Schwinge  <thomas@codesourcery.com>
+
+	* c-parser.c (c_parser_oacc_clause_use_device): Merge function
+	into...
+	(c_parser_omp_clause_use_device_ptr): ... this function.  Adjust
+	all users.
+
 2015-12-22  Marek Polacek  <polacek@redhat.com>
 
 	PR c/69002
diff --git gcc/c/c-parser.c gcc/c/c-parser.c
index 353e3da..8e754d0 100644
--- gcc/c/c-parser.c
+++ gcc/c/c-parser.c
@@ -11395,7 +11395,10 @@  c_parser_omp_clause_defaultmap (c_parser *parser, tree list)
   return list;
 }
 
-/* OpenMP 4.5:
+/* OpenACC 2.0:
+   use_device ( variable-list )
+
+   OpenMP 4.5:
    use_device_ptr ( variable-list ) */
 
 static tree
@@ -11730,15 +11733,6 @@  c_parser_oacc_clause_tile (c_parser *parser, tree list)
   return c;
 }
 
-/* OpenACC 2.0:
-   use_device ( variable-list ) */
-
-static tree
-c_parser_oacc_clause_use_device (c_parser *parser, tree list)
-{
-  return c_parser_omp_var_list_parens (parser, OMP_CLAUSE_USE_DEVICE, list);
-}
-
 /* OpenACC:
    wait ( int-expr-list ) */
 
@@ -13058,7 +13052,7 @@  c_parser_oacc_all_clauses (c_parser *parser, omp_clause_mask mask,
 	  c_name = "tile";
 	  break;
 	case PRAGMA_OACC_CLAUSE_USE_DEVICE:
-	  clauses = c_parser_oacc_clause_use_device (parser, clauses);
+	  clauses = c_parser_omp_clause_use_device_ptr (parser, clauses);
 	  c_name = "use_device";
 	  break;
 	case PRAGMA_OACC_CLAUSE_VECTOR:
diff --git gcc/c/c-typeck.c gcc/c/c-typeck.c
index 928fcd5..7406bd4 100644
--- gcc/c/c-typeck.c
+++ gcc/c/c-typeck.c
@@ -13125,7 +13125,6 @@  c_finish_omp_clauses (tree clauses, bool is_omp, bool declare_simd)
 	  bitmap_set_bit (&map_head, DECL_UID (t));
 	  goto check_dup_generic;
 
-	case OMP_CLAUSE_USE_DEVICE:
 	case OMP_CLAUSE_IS_DEVICE_PTR:
 	case OMP_CLAUSE_USE_DEVICE_PTR:
 	  t = OMP_CLAUSE_DECL (c);
diff --git gcc/cp/parser.c gcc/cp/parser.c
index 842dded..4829a77 100644
--- gcc/cp/parser.c
+++ gcc/cp/parser.c
@@ -32097,7 +32097,7 @@  cp_parser_oacc_all_clauses (cp_parser *parser, omp_clause_mask mask,
 	  c_name = "tile";
 	  break;
 	case PRAGMA_OACC_CLAUSE_USE_DEVICE:
-	  clauses = cp_parser_omp_var_list (parser, OMP_CLAUSE_USE_DEVICE,
+	  clauses = cp_parser_omp_var_list (parser, OMP_CLAUSE_USE_DEVICE_PTR,
 					    clauses);
 	  c_name = "use_device";
 	  break;
diff --git gcc/cp/pt.c gcc/cp/pt.c
index dab15bd..4555b32 100644
--- gcc/cp/pt.c
+++ gcc/cp/pt.c
@@ -14425,7 +14425,6 @@  tsubst_omp_clauses (tree clauses, bool declare_simd, bool allow_fields,
 	case OMP_CLAUSE_FROM:
 	case OMP_CLAUSE_TO:
 	case OMP_CLAUSE_MAP:
-	case OMP_CLAUSE_USE_DEVICE:
 	case OMP_CLAUSE_USE_DEVICE_PTR:
 	case OMP_CLAUSE_IS_DEVICE_PTR:
 	  OMP_CLAUSE_DECL (nc)
@@ -14552,7 +14551,6 @@  tsubst_omp_clauses (tree clauses, bool declare_simd, bool allow_fields,
 	  case OMP_CLAUSE_COPYPRIVATE:
 	  case OMP_CLAUSE_LINEAR:
 	  case OMP_CLAUSE_REDUCTION:
-	  case OMP_CLAUSE_USE_DEVICE:
 	  case OMP_CLAUSE_USE_DEVICE_PTR:
 	  case OMP_CLAUSE_IS_DEVICE_PTR:
 	    /* tsubst_expr on SCOPE_REF results in returning
diff --git gcc/cp/semantics.c gcc/cp/semantics.c
index ab9989a..37bf050 100644
--- gcc/cp/semantics.c
+++ gcc/cp/semantics.c
@@ -6886,7 +6886,6 @@  finish_omp_clauses (tree clauses, bool allow_fields, bool declare_simd)
 	    }
 	  break;
 
-	case OMP_CLAUSE_USE_DEVICE:
 	case OMP_CLAUSE_IS_DEVICE_PTR:
 	case OMP_CLAUSE_USE_DEVICE_PTR:
 	  field_ok = allow_fields;
diff --git gcc/fortran/trans-openmp.c gcc/fortran/trans-openmp.c
index 227964c..70a7722 100644
--- gcc/fortran/trans-openmp.c
+++ gcc/fortran/trans-openmp.c
@@ -1771,7 +1771,7 @@  gfc_trans_omp_clauses (stmtblock_t *block, gfc_omp_clauses *clauses,
 	  clause_code = OMP_CLAUSE_UNIFORM;
 	  goto add_clause;
 	case OMP_LIST_USE_DEVICE:
-	  clause_code = OMP_CLAUSE_USE_DEVICE;
+	  clause_code = OMP_CLAUSE_USE_DEVICE_PTR;
 	  goto add_clause;
 	case OMP_LIST_DEVICE_RESIDENT:
 	  clause_code = OMP_CLAUSE_DEVICE_RESIDENT;
diff --git gcc/gimplify.c gcc/gimplify.c
index 62b0e64..bc90401 100644
--- gcc/gimplify.c
+++ gcc/gimplify.c
@@ -7139,7 +7139,6 @@  gimplify_scan_omp_clauses (tree *list_p, gimple_seq *pre_p,
 	    }
 	  goto do_notice;
 
-	case OMP_CLAUSE_USE_DEVICE:
 	case OMP_CLAUSE_USE_DEVICE_PTR:
 	  flags = GOVD_FIRSTPRIVATE | GOVD_EXPLICIT;
 	  goto do_add;
@@ -8051,7 +8050,6 @@  gimplify_adjust_omp_clauses (gimple_seq *pre_p, gimple_seq body, tree *list_p,
 	case OMP_CLAUSE_ASYNC:
 	case OMP_CLAUSE_WAIT:
 	case OMP_CLAUSE_DEVICE_RESIDENT:
-	case OMP_CLAUSE_USE_DEVICE:
 	case OMP_CLAUSE_INDEPENDENT:
 	case OMP_CLAUSE_NUM_GANGS:
 	case OMP_CLAUSE_NUM_WORKERS:
diff --git gcc/omp-low.c gcc/omp-low.c
index 676b1df..a0c3e1c 100644
--- gcc/omp-low.c
+++ gcc/omp-low.c
@@ -1957,7 +1957,6 @@  scan_sharing_clauses (tree clauses, omp_context *ctx,
 	  install_var_local (decl, ctx);
 	  break;
 
-	case OMP_CLAUSE_USE_DEVICE:
 	case OMP_CLAUSE_USE_DEVICE_PTR:
 	  decl = OMP_CLAUSE_DECL (c);
 	  if (TREE_CODE (TREE_TYPE (decl)) == ARRAY_TYPE)
@@ -2314,7 +2313,6 @@  scan_sharing_clauses (tree clauses, omp_context *ctx,
 	case OMP_CLAUSE_SIMD:
 	case OMP_CLAUSE_NOGROUP:
 	case OMP_CLAUSE_DEFAULTMAP:
-	case OMP_CLAUSE_USE_DEVICE:
 	case OMP_CLAUSE_USE_DEVICE_PTR:
 	case OMP_CLAUSE__CILK_FOR_COUNT_:
 	case OMP_CLAUSE_ASYNC:
@@ -15288,7 +15286,6 @@  lower_omp_target (gimple_stmt_iterator *gsi_p, omp_context *ctx)
 	  }
 	break;
 
-      case OMP_CLAUSE_USE_DEVICE:
       case OMP_CLAUSE_USE_DEVICE_PTR:
       case OMP_CLAUSE_IS_DEVICE_PTR:
 	var = OMP_CLAUSE_DECL (c);
@@ -15674,14 +15671,12 @@  lower_omp_target (gimple_stmt_iterator *gsi_p, omp_context *ctx)
 				    build_int_cstu (tkind_type, tkind));
 	    break;
 
-	  case OMP_CLAUSE_USE_DEVICE:
 	  case OMP_CLAUSE_USE_DEVICE_PTR:
 	  case OMP_CLAUSE_IS_DEVICE_PTR:
 	    ovar = OMP_CLAUSE_DECL (c);
 	    var = lookup_decl_in_outer_ctx (ovar, ctx);
 	    x = build_sender_ref (ovar, ctx);
-	    if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_USE_DEVICE_PTR
-		|| OMP_CLAUSE_CODE (c) == OMP_CLAUSE_USE_DEVICE)
+	    if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_USE_DEVICE_PTR)
 	      tkind = GOMP_MAP_USE_DEVICE_PTR;
 	    else
 	      tkind = GOMP_MAP_FIRSTPRIVATE_INT;
@@ -15884,12 +15879,10 @@  lower_omp_target (gimple_stmt_iterator *gsi_p, omp_context *ctx)
 				     gimple_build_assign (new_var, x));
 	      }
 	    break;
-	  case OMP_CLAUSE_USE_DEVICE:
 	  case OMP_CLAUSE_USE_DEVICE_PTR:
 	  case OMP_CLAUSE_IS_DEVICE_PTR:
 	    var = OMP_CLAUSE_DECL (c);
-	    if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_USE_DEVICE_PTR
-		|| OMP_CLAUSE_CODE (c) == OMP_CLAUSE_USE_DEVICE)
+	    if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_USE_DEVICE_PTR)
 	      x = build_sender_ref (var, ctx);
 	    else
 	      x = build_receiver_ref (var, false, ctx);
diff --git gcc/testsuite/gfortran.dg/goacc/host_data-tree.f95 gcc/testsuite/gfortran.dg/goacc/host_data-tree.f95
index 7a5eea6..23aba8c 100644
--- gcc/testsuite/gfortran.dg/goacc/host_data-tree.f95
+++ gcc/testsuite/gfortran.dg/goacc/host_data-tree.f95
@@ -8,4 +8,4 @@  program test
   !$acc host_data use_device(i)
   !$acc end host_data
 end program test
-! { dg-final { scan-tree-dump-times "pragma acc host_data use_device\\(i\\)" 1 "original" } } 
+! { dg-final { scan-tree-dump-times "pragma acc host_data use_device_ptr\\(i\\)" 1 "original" } }
diff --git gcc/tree-core.h gcc/tree-core.h
index 9cc64d9..5371378 100644
--- gcc/tree-core.h
+++ gcc/tree-core.h
@@ -302,7 +302,8 @@  enum omp_clause_code {
      OpenMP clause: map ({alloc:,to:,from:,tofrom:,}variable-list).  */
   OMP_CLAUSE_MAP,
 
-  /* OpenMP clause: use_device_ptr (variable-list).  */
+  /* OpenACC clause: use_device (variable_list).
+     OpenMP clause: use_device_ptr (variable-list).  */
   OMP_CLAUSE_USE_DEVICE_PTR,
 
   /* OpenMP clause: is_device_ptr (variable-list).  */
@@ -315,9 +316,6 @@  enum omp_clause_code {
   /* OpenACC clause: device_resident (variable_list).  */
   OMP_CLAUSE_DEVICE_RESIDENT,
 
-  /* OpenACC clause: use_device (variable_list).  */
-  OMP_CLAUSE_USE_DEVICE,
-
   /* OpenACC clause: gang [(gang-argument-list)].
      Where
       gang-argument-list: [gang-argument-list, ] gang-argument
diff --git gcc/tree-nested.c gcc/tree-nested.c
index 3a9479a..8211a12 100644
--- gcc/tree-nested.c
+++ gcc/tree-nested.c
@@ -1072,7 +1072,6 @@  convert_nonlocal_omp_clauses (tree *pclauses, struct walk_stmt_info *wi)
 	case OMP_CLAUSE_SHARED:
 	case OMP_CLAUSE_TO_DECLARE:
 	case OMP_CLAUSE_LINK:
-	case OMP_CLAUSE_USE_DEVICE:
 	case OMP_CLAUSE_USE_DEVICE_PTR:
 	case OMP_CLAUSE_IS_DEVICE_PTR:
 	do_decl_clause:
@@ -1744,7 +1743,6 @@  convert_local_omp_clauses (tree *pclauses, struct walk_stmt_info *wi)
 	case OMP_CLAUSE_SHARED:
 	case OMP_CLAUSE_TO_DECLARE:
 	case OMP_CLAUSE_LINK:
-	case OMP_CLAUSE_USE_DEVICE:
 	case OMP_CLAUSE_USE_DEVICE_PTR:
 	case OMP_CLAUSE_IS_DEVICE_PTR:
 	do_decl_clause:
diff --git gcc/tree-pretty-print.c gcc/tree-pretty-print.c
index caec760..c4a180e 100644
--- gcc/tree-pretty-print.c
+++ gcc/tree-pretty-print.c
@@ -327,9 +327,6 @@  dump_omp_clause (pretty_printer *pp, tree clause, int spc, int flags)
     case OMP_CLAUSE_DEVICE_RESIDENT:
       name = "device_resident";
       goto print_remap;
-    case OMP_CLAUSE_USE_DEVICE:
-      name = "use_device";
-      goto print_remap;
     case OMP_CLAUSE_TO_DECLARE:
       name = "to";
       goto print_remap;
diff --git gcc/tree.c gcc/tree.c
index 2190cae..d837609 100644
--- gcc/tree.c
+++ gcc/tree.c
@@ -282,7 +282,6 @@  unsigned const char omp_clause_num_ops[] =
   1, /* OMP_CLAUSE_IS_DEVICE_PTR  */
   2, /* OMP_CLAUSE__CACHE_  */
   1, /* OMP_CLAUSE_DEVICE_RESIDENT  */
-  1, /* OMP_CLAUSE_USE_DEVICE  */
   2, /* OMP_CLAUSE_GANG  */
   1, /* OMP_CLAUSE_ASYNC  */
   1, /* OMP_CLAUSE_WAIT  */
@@ -354,7 +353,6 @@  const char * const omp_clause_code_name[] =
   "is_device_ptr",
   "_cache_",
   "device_resident",
-  "use_device",
   "gang",
   "async",
   "wait",
@@ -11612,7 +11610,6 @@  walk_tree_1 (tree *tp, walk_tree_fn func, void *data,
 	  /* FALLTHRU */
 
 	case OMP_CLAUSE_DEVICE_RESIDENT:
-	case OMP_CLAUSE_USE_DEVICE:
 	case OMP_CLAUSE_ASYNC:
 	case OMP_CLAUSE_WAIT:
 	case OMP_CLAUSE_WORKER: