tree-core.h (enum omp_clause_code): Adjust OMP_CLAUSE_USE_DEVICE_PTR OpenMP description.

	* tree-core.h (enum omp_clause_code): Adjust OMP_CLAUSE_USE_DEVICE_PTR
	OpenMP description.  Add OMP_CLAUSE_USE_DEVICE_ADDR clause.
	* tree.c (omp_clause_num_ops, omp_clause_code_name): Add entries
	for OMP_CLAUSE_USE_DEVICE_ADDR clause.
	(walk_tree_1): Handle OMP_CLAUSE_USE_DEVICE_ADDR.
	* tree-pretty-print.c (dump_omp_clause): Likewise.
	* tree-nested.c (convert_nonlocal_omp_clauses,
	convert_local_omp_clauses): Likewise.
	* gimplify.c (gimplify_scan_omp_clauses, gimplify_adjust_omp_clauses):
	Likewise.
	* omp-low.c (scan_sharing_clauses, lower_omp_target): Likewise.
	Treat OMP_CLAUSE_USE_DEVICE_ADDR like OMP_CLAUSE_USE_DEVICE_PTR
	clause with array or reference to array types, no matter what type
	except for reference it has.
gcc/c-family/
	* c-pragma.h (enum pragma_omp_clause): Add
	PRAGMA_OMP_CLAUSE_USE_DEVICE_ADDR.  Set PRAGMA_OACC_CLAUSE_USE_DEVICE
	equal to PRAGMA_OMP_CLAUSE_USE_DEVICE_PTR instead of being a separate
	enumeration value.
gcc/c/
	* c-parser.c (c_parser_omp_clause_name): Parse use_device_addr clause.
	(c_parser_omp_clause_use_device_addr): New function.
	(c_parser_omp_all_clauses): Handle PRAGMA_OMP_CLAUSE_USE_DEVICE_ADDR.
	(OMP_TARGET_DATA_CLAUSE_MASK): Add PRAGMA_OMP_CLAUSE_USE_DEVICE_ADDR.
	(c_parser_omp_target_data): Handle PRAGMA_OMP_CLAUSE_USE_DEVICE_ADDR
	like PRAGMA_OMP_CLAUSE_USE_DEVICE_PTR, adjust diagnostics about no
	map or use_device_* clauses.
	* c-typeck.c (c_finish_omp_clauses): For OMP_CLAUSE_USE_DEVICE_PTR
	in OpenMP, require pointer type rather than pointer or array type.
	Handle OMP_CLAUSE_USE_DEVICE_ADDR.
gcc/cp/
	* parser.c (cp_parser_omp_clause_name): Parse use_device_addr clause.
	(cp_parser_omp_all_clauses): Handle PRAGMA_OMP_CLAUSE_USE_DEVICE_ADDR.
	(OMP_TARGET_DATA_CLAUSE_MASK): Add PRAGMA_OMP_CLAUSE_USE_DEVICE_ADDR.
	(cp_parser_omp_target_data): Handle PRAGMA_OMP_CLAUSE_USE_DEVICE_ADDR
	like PRAGMA_OMP_CLAUSE_USE_DEVICE_PTR, adjust diagnostics about no
	map or use_device_* clauses.
	* semantics.c (finish_omp_clauses): For OMP_CLAUSE_USE_DEVICE_PTR
	in OpenMP, require pointer or reference to pointer type rather than
	pointer or array or reference to pointer or array type. Handle
	OMP_CLAUSE_USE_DEVICE_ADDR.
	* pt.c (tsubst_omp_clauses): Handle OMP_CLAUSE_USE_DEVICE_ADDR.
gcc/testsuite/
	* c-c++-common/gomp/target-data-1.c (foo): Use use_device_addr clause
	instead of use_device_ptr clause where required by OpenMP 5.0, add
	further tests for both use_device_ptr and use_device_addr clauses.
libgomp/
	* testsuite/libgomp.c/target-18.c (struct S): New type.
	(foo): Use use_device_addr clause instead of use_device_ptr clause
	where required by OpenMP 5.0, add further tests for both use_device_ptr
	and use_device_addr clauses.
	* testsuite/libgomp.c++/target-9.C (struct S): New type.
	(foo): Use use_device_addr clause instead of use_device_ptr clause
	where required by OpenMP 5.0, add further tests for both use_device_ptr
	and use_device_addr clauses.  Add t and u arguments.
	(main): Adjust caller.

From-SVN: r274159

1 files changed, 22 insertions, 3 deletions

diff --git a/gcc/c/c-parser.c b/gcc/c/c-parser.c
index c7c13a6..9b5cfcf 100644
--- a/gcc/c/c-parser.c
+++ b/gcc/c/c-parser.c
@@ -11866,6 +11866,8 @@ c_parser_omp_clause_name (c_parser *parser)
 	    result = PRAGMA_OMP_CLAUSE_UNTIED;
 	  else if (!strcmp ("use_device", p))
 	    result = PRAGMA_OACC_CLAUSE_USE_DEVICE;
+	  else if (!strcmp ("use_device_addr", p))
+	    result = PRAGMA_OMP_CLAUSE_USE_DEVICE_ADDR;
 	  else if (!strcmp ("use_device_ptr", p))
 	    result = PRAGMA_OMP_CLAUSE_USE_DEVICE_PTR;
 	  break;
@@ -13121,6 +13123,16 @@ c_parser_omp_clause_use_device_ptr (c_parser *parser, tree list)
 				       list);
 }
 
+/* OpenMP 5.0:
+   use_device_addr ( variable-list ) */
+
+static tree
+c_parser_omp_clause_use_device_addr (c_parser *parser, tree list)
+{
+  return c_parser_omp_var_list_parens (parser, OMP_CLAUSE_USE_DEVICE_ADDR,
+				       list);
+}
+
 /* OpenMP 4.5:
    is_device_ptr ( variable-list ) */
 
@@ -15321,6 +15333,10 @@ c_parser_omp_all_clauses (c_parser *parser, omp_clause_mask mask,
 	  clauses = c_parser_omp_clause_use_device_ptr (parser, clauses);
 	  c_name = "use_device_ptr";
 	  break;
+	case PRAGMA_OMP_CLAUSE_USE_DEVICE_ADDR:
+	  clauses = c_parser_omp_clause_use_device_addr (parser, clauses);
+	  c_name = "use_device_addr";
+	  break;
 	case PRAGMA_OMP_CLAUSE_IS_DEVICE_PTR:
 	  clauses = c_parser_omp_clause_is_device_ptr (parser, clauses);
 	  c_name = "is_device_ptr";
@@ -18288,7 +18304,8 @@ c_parser_omp_teams (location_t loc, c_parser *parser,
 	( (OMP_CLAUSE_MASK_1 << PRAGMA_OMP_CLAUSE_DEVICE)	\
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OMP_CLAUSE_MAP)		\
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OMP_CLAUSE_IF)		\
-	| (OMP_CLAUSE_MASK_1 << PRAGMA_OMP_CLAUSE_USE_DEVICE_PTR))
+	| (OMP_CLAUSE_MASK_1 << PRAGMA_OMP_CLAUSE_USE_DEVICE_PTR) \
+	| (OMP_CLAUSE_MASK_1 << PRAGMA_OMP_CLAUSE_USE_DEVICE_ADDR))
 
 static tree
 c_parser_omp_target_data (location_t loc, c_parser *parser, bool *if_p)
@@ -18323,7 +18340,8 @@ c_parser_omp_target_data (location_t loc, c_parser *parser, bool *if_p)
 	    *pc = OMP_CLAUSE_CHAIN (*pc);
 	    continue;
 	  }
-      else if (OMP_CLAUSE_CODE (*pc) == OMP_CLAUSE_USE_DEVICE_PTR)
+      else if (OMP_CLAUSE_CODE (*pc) == OMP_CLAUSE_USE_DEVICE_PTR
+	       || OMP_CLAUSE_CODE (*pc) == OMP_CLAUSE_USE_DEVICE_ADDR)
 	map_seen = 3;
       pc = &OMP_CLAUSE_CHAIN (*pc);
     }
@@ -18333,7 +18351,8 @@ c_parser_omp_target_data (location_t loc, c_parser *parser, bool *if_p)
       if (map_seen == 0)
 	error_at (loc,
 		  "%<#pragma omp target data%> must contain at least "
-		  "one %<map%> or %<use_device_ptr%> clause");
+		  "one %<map%>, %<use_device_ptr%> or %<use_device_addr%> "
+		  "clause");
       return NULL_TREE;
     }