@@ -13084,6 +13084,175 @@ c_check_omp_declare_reduction_r (tree *tp, int *, void *data)
return NULL_TREE;
}
+/* Return identifier to look up for omp declare reduction. */
+
+tree
+c_omp_mapper_id (tree mapper_id)
+{
+ const char *p = NULL;
+
+ const char prefix[] = "omp declare mapper ";
+
+ if (mapper_id == NULL_TREE)
+ p = "<default>";
+ else if (TREE_CODE (mapper_id) == IDENTIFIER_NODE)
+ p = IDENTIFIER_POINTER (mapper_id);
+ else
+ return error_mark_node;
+
+ size_t lenp = sizeof (prefix);
+ size_t len = strlen (p);
+ char *name = XALLOCAVEC (char, lenp + len);
+ memcpy (name, prefix, lenp - 1);
+ memcpy (name + lenp - 1, p, len + 1);
+ return get_identifier (name);
+}
+
+/* Lookup MAPPER_ID in the current scope, or create an artificial
+ VAR_DECL, bind it into the current scope and return it. */
+
+tree
+c_omp_mapper_decl (tree mapper_id)
+{
+ struct c_binding *b = I_SYMBOL_BINDING (mapper_id);
+ if (b != NULL && B_IN_CURRENT_SCOPE (b))
+ return b->decl;
+
+ tree decl = build_decl (BUILTINS_LOCATION, VAR_DECL,
+ mapper_id, integer_type_node);
+ DECL_ARTIFICIAL (decl) = 1;
+ DECL_EXTERNAL (decl) = 1;
+ TREE_STATIC (decl) = 1;
+ TREE_PUBLIC (decl) = 0;
+ bind (mapper_id, decl, current_scope, true, false, BUILTINS_LOCATION);
+ return decl;
+}
+
+/* Lookup MAPPER_ID in the first scope where it has entry for TYPE. */
+
+tree
+c_omp_mapper_lookup (tree mapper_id, tree type)
+{
+ if (TREE_CODE (type) != RECORD_TYPE
+ && TREE_CODE (type) != UNION_TYPE)
+ return NULL_TREE;
+
+ mapper_id = c_omp_mapper_id (mapper_id);
+
+ struct c_binding *b = I_SYMBOL_BINDING (mapper_id);
+ while (b)
+ {
+ tree t;
+ for (t = DECL_INITIAL (b->decl); t; t = TREE_CHAIN (t))
+ if (comptypes (TREE_PURPOSE (t), type))
+ return TREE_VALUE (t);
+ b = b->shadowed;
+ }
+ return NULL_TREE;
+}
+
+/* For C, we record a pointer to the mapper itself without wrapping it in an
+ artificial function or similar. So, just return it. */
+
+tree
+c_omp_extract_mapper_directive (tree mapper)
+{
+ return mapper;
+}
+
+/* For now we can handle singleton OMP_ARRAY_SECTIONs with custom mappers, but
+ nothing more complicated. */
+
+tree
+c_omp_map_array_section (location_t loc, tree t)
+{
+ tree low = TREE_OPERAND (t, 1);
+ tree len = TREE_OPERAND (t, 2);
+
+ if (len && integer_onep (len))
+ {
+ t = TREE_OPERAND (t, 0);
+
+ if (!low)
+ low = integer_zero_node;
+
+ t = build_array_ref (loc, t, low);
+ }
+
+ return t;
+}
+
+/* Helper function for below function. */
+
+static tree
+c_omp_scan_mapper_bindings_r (tree *tp, int *walk_subtrees, void *ptr)
+{
+ tree t = *tp;
+ omp_mapper_list<tree> *mlist = (omp_mapper_list<tree> *) ptr;
+ tree aggr_type = NULL_TREE;
+
+ if (TREE_CODE (t) == SIZEOF_EXPR
+ || TREE_CODE (t) == ALIGNOF_EXPR)
+ {
+ *walk_subtrees = 0;
+ return NULL_TREE;
+ }
+
+ if (TREE_CODE (t) == OMP_CLAUSE)
+ return NULL_TREE;
+
+ if (TREE_CODE (t) == COMPONENT_REF
+ && AGGREGATE_TYPE_P (TREE_TYPE (TREE_OPERAND (t, 0))))
+ aggr_type = TREE_TYPE (TREE_OPERAND (t, 0));
+ else if ((TREE_CODE (t) == VAR_DECL
+ || TREE_CODE (t) == PARM_DECL
+ || TREE_CODE (t) == RESULT_DECL)
+ && AGGREGATE_TYPE_P (TREE_TYPE (t)))
+ aggr_type = TREE_TYPE (t);
+
+ if (aggr_type)
+ {
+ tree mapper_fn = c_omp_mapper_lookup (NULL_TREE, aggr_type);
+ if (mapper_fn)
+ mlist->add_mapper (NULL_TREE, aggr_type, mapper_fn);
+ }
+
+ return NULL_TREE;
+}
+
+/* Scan an offload region's body, and record uses of struct- or union-typed
+ variables. Add _mapper_binding_ fake clauses to *CLAUSES_PTR. */
+
+void
+c_omp_scan_mapper_bindings (location_t loc, tree *clauses_ptr, tree body)
+{
+ hash_set<omp_name_type<tree>> seen_types;
+ auto_vec<tree> mappers;
+ omp_mapper_list<tree> mlist (&seen_types, &mappers);
+
+ walk_tree_without_duplicates (&body, c_omp_scan_mapper_bindings_r, &mlist);
+
+ unsigned int i;
+ tree mapper;
+ FOR_EACH_VEC_ELT (mappers, i, mapper)
+ c_omp_find_nested_mappers (&mlist, mapper);
+
+ FOR_EACH_VEC_ELT (mappers, i, mapper)
+ {
+ if (mapper == error_mark_node)
+ continue;
+ tree mapper_name = OMP_DECLARE_MAPPER_ID (mapper);
+ tree decl = OMP_DECLARE_MAPPER_DECL (mapper);
+
+ tree c = build_omp_clause (loc, OMP_CLAUSE__MAPPER_BINDING_);
+ OMP_CLAUSE__MAPPER_BINDING__ID (c) = mapper_name;
+ OMP_CLAUSE__MAPPER_BINDING__DECL (c) = decl;
+ OMP_CLAUSE__MAPPER_BINDING__MAPPER (c) = mapper;
+
+ OMP_CLAUSE_CHAIN (c) = *clauses_ptr;
+ *clauses_ptr = c;
+ }
+}
bool
c_check_in_current_scope (tree decl)
@@ -122,6 +122,18 @@ along with GCC; see the file COPYING3. If not see
#undef LANG_HOOKS_OMP_CLAUSE_ASSIGN_OP
#define LANG_HOOKS_OMP_CLAUSE_ASSIGN_OP c_omp_clause_copy_ctor
+#undef LANG_HOOKS_OMP_FINISH_MAPPER_CLAUSES
+#define LANG_HOOKS_OMP_FINISH_MAPPER_CLAUSES c_omp_finish_mapper_clauses
+
+#undef LANG_HOOKS_OMP_MAPPER_LOOKUP
+#define LANG_HOOKS_OMP_MAPPER_LOOKUP c_omp_mapper_lookup
+
+#undef LANG_HOOKS_OMP_EXTRACT_MAPPER_DIRECTIVE
+#define LANG_HOOKS_OMP_EXTRACT_MAPPER_DIRECTIVE c_omp_extract_mapper_directive
+
+#undef LANG_HOOKS_OMP_MAP_ARRAY_SECTION
+#define LANG_HOOKS_OMP_MAP_ARRAY_SECTION c_omp_map_array_section
+
#undef LANG_HOOKS_TREE_INLINING_VAR_MOD_TYPE_P
#define LANG_HOOKS_TREE_INLINING_VAR_MOD_TYPE_P c_vla_unspec_p
#endif /* GCC_C_OBJC_COMMON */
@@ -17221,10 +17221,9 @@ c_parser_omp_clause_doacross (c_parser *parser, tree list)
always | close */
static tree
-c_parser_omp_clause_map (c_parser *parser, tree list)
+c_parser_omp_clause_map (c_parser *parser, tree list, enum gomp_map_kind kind)
{
location_t clause_loc = c_parser_peek_token (parser)->location;
- enum gomp_map_kind kind = GOMP_MAP_TOFROM;
tree nl, c;
matching_parens parens;
@@ -17243,11 +17242,27 @@ c_parser_omp_clause_map (c_parser *parser, tree list)
if (c_parser_peek_nth_token_raw (parser, pos + 1)->type == CPP_COMMA)
pos++;
+ else if ((c_parser_peek_nth_token_raw (parser, pos + 1)->type
+ == CPP_OPEN_PAREN)
+ && ((c_parser_peek_nth_token_raw (parser, pos + 2)->type
+ == CPP_NAME)
+ || ((c_parser_peek_nth_token_raw (parser, pos + 2)->type
+ == CPP_KEYWORD)
+ && (c_parser_peek_nth_token_raw (parser,
+ pos + 2)->keyword
+ == RID_DEFAULT)))
+ && (c_parser_peek_nth_token_raw (parser, pos + 3)->type
+ == CPP_CLOSE_PAREN)
+ && (c_parser_peek_nth_token_raw (parser, pos + 4)->type
+ == CPP_COMMA))
+ pos += 4;
pos++;
}
int always_modifier = 0;
int close_modifier = 0;
+ int mapper_modifier = 0;
+ tree mapper_name = NULL_TREE;
for (int pos = 1; pos < map_kind_pos; ++pos)
{
c_token *tok = c_parser_peek_token (parser);
@@ -17268,6 +17283,7 @@ c_parser_omp_clause_map (c_parser *parser, tree list)
return list;
}
always_modifier++;
+ c_parser_consume_token (parser);
}
else if (strcmp ("close", p) == 0)
{
@@ -17278,6 +17294,60 @@ c_parser_omp_clause_map (c_parser *parser, tree list)
return list;
}
close_modifier++;
+ c_parser_consume_token (parser);
+ }
+ else if (strcmp ("mapper", p) == 0)
+ {
+ c_parser_consume_token (parser);
+
+ matching_parens mparens;
+ if (mparens.require_open (parser))
+ {
+ if (mapper_modifier)
+ {
+ c_parser_error (parser, "too many %<mapper%> modifiers");
+ /* Assume it's a well-formed mapper modifier, even if it
+ seems to be in the wrong place. */
+ c_parser_consume_token (parser);
+ mparens.require_close (parser);
+ parens.skip_until_found_close (parser);
+ return list;
+ }
+
+ tok = c_parser_peek_token (parser);
+
+ switch (tok->type)
+ {
+ case CPP_NAME:
+ {
+ mapper_name = tok->value;
+ c_parser_consume_token (parser);
+ }
+ break;
+
+ case CPP_KEYWORD:
+ if (tok->keyword == RID_DEFAULT)
+ {
+ c_parser_consume_token (parser);
+ break;
+ }
+ /* Fallthrough. */
+
+ default:
+ error_at (tok->location,
+ "expected identifier or %<default%>");
+ return list;
+ }
+
+ if (!mparens.require_close (parser))
+ {
+ parens.skip_until_found_close (parser);
+ return list;
+ }
+
+ mapper_modifier++;
+ pos += 3;
+ }
}
else
{
@@ -17287,8 +17357,6 @@ c_parser_omp_clause_map (c_parser *parser, tree list)
parens.skip_until_found_close (parser);
return list;
}
-
- c_parser_consume_token (parser);
}
if (c_parser_next_token_is (parser, CPP_NAME)
@@ -17321,8 +17389,30 @@ c_parser_omp_clause_map (c_parser *parser, tree list)
nl = c_parser_omp_variable_list (parser, clause_loc, OMP_CLAUSE_MAP, list,
true);
+ tree last_new = NULL_TREE;
+
for (c = nl; c != list; c = OMP_CLAUSE_CHAIN (c))
- OMP_CLAUSE_SET_MAP_KIND (c, kind);
+ {
+ OMP_CLAUSE_SET_MAP_KIND (c, kind);
+ last_new = c;
+ }
+
+ if (mapper_name)
+ {
+ tree name = build_omp_clause (input_location, OMP_CLAUSE_MAP);
+ OMP_CLAUSE_SET_MAP_KIND (name, GOMP_MAP_PUSH_MAPPER_NAME);
+ OMP_CLAUSE_DECL (name) = mapper_name;
+ OMP_CLAUSE_CHAIN (name) = nl;
+ nl = name;
+
+ gcc_assert (last_new);
+
+ name = build_omp_clause (input_location, OMP_CLAUSE_MAP);
+ OMP_CLAUSE_SET_MAP_KIND (name, GOMP_MAP_POP_MAPPER_NAME);
+ OMP_CLAUSE_DECL (name) = null_pointer_node;
+ OMP_CLAUSE_CHAIN (name) = OMP_CLAUSE_CHAIN (last_new);
+ OMP_CLAUSE_CHAIN (last_new) = name;
+ }
parens.skip_until_found_close (parser);
return nl;
@@ -18125,7 +18215,7 @@ c_parser_omp_all_clauses (c_parser *parser, omp_clause_mask mask,
c_name = "doacross";
break;
case PRAGMA_OMP_CLAUSE_MAP:
- clauses = c_parser_omp_clause_map (parser, clauses);
+ clauses = c_parser_omp_clause_map (parser, clauses, GOMP_MAP_TOFROM);
c_name = "map";
break;
case PRAGMA_OMP_CLAUSE_USE_DEVICE_PTR:
@@ -22169,7 +22259,7 @@ c_parser_omp_target (c_parser *parser, enum pragma_context context, bool *if_p)
{
location_t loc = c_parser_peek_token (parser)->location;
c_parser_consume_pragma (parser);
- tree *pc = NULL, stmt, block;
+ tree *pc = NULL, stmt, block, body, clauses;
if (context != pragma_stmt && context != pragma_compound)
{
@@ -22324,10 +22414,9 @@ c_parser_omp_target (c_parser *parser, enum pragma_context context, bool *if_p)
stmt = make_node (OMP_TARGET);
TREE_TYPE (stmt) = void_type_node;
- OMP_TARGET_CLAUSES (stmt)
- = c_parser_omp_all_clauses (parser, OMP_TARGET_CLAUSE_MASK,
- "#pragma omp target", false);
- for (tree c = OMP_TARGET_CLAUSES (stmt); c; c = OMP_CLAUSE_CHAIN (c))
+ clauses = c_parser_omp_all_clauses (parser, OMP_TARGET_CLAUSE_MASK,
+ "#pragma omp target", false);
+ for (tree c = clauses; c; c = OMP_CLAUSE_CHAIN (c))
if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_IN_REDUCTION)
{
tree nc = build_omp_clause (OMP_CLAUSE_LOCATION (c), OMP_CLAUSE_MAP);
@@ -22336,14 +22425,19 @@ c_parser_omp_target (c_parser *parser, enum pragma_context context, bool *if_p)
OMP_CLAUSE_CHAIN (nc) = OMP_CLAUSE_CHAIN (c);
OMP_CLAUSE_CHAIN (c) = nc;
}
- OMP_TARGET_CLAUSES (stmt)
- = c_finish_omp_clauses (OMP_TARGET_CLAUSES (stmt), C_ORT_OMP_TARGET);
- c_omp_adjust_map_clauses (OMP_TARGET_CLAUSES (stmt), true);
+ clauses = c_omp_instantiate_mappers (clauses);
+ clauses = c_finish_omp_clauses (clauses, C_ORT_OMP_TARGET);
+ c_omp_adjust_map_clauses (clauses, true);
- pc = &OMP_TARGET_CLAUSES (stmt);
keep_next_level ();
block = c_begin_compound_stmt (true);
- add_stmt (c_parser_omp_structured_block (parser, if_p));
+ body = c_parser_omp_structured_block (parser, if_p);
+
+ c_omp_scan_mapper_bindings (loc, &clauses, body);
+
+ add_stmt (body);
+ OMP_TARGET_CLAUSES (stmt) = clauses;
+ pc = &OMP_TARGET_CLAUSES (stmt);
OMP_TARGET_BODY (stmt) = c_end_compound_stmt (loc, block, true);
SET_EXPR_LOCATION (stmt, loc);
@@ -23644,6 +23738,151 @@ c_parser_omp_declare_reduction (c_parser *parser, enum pragma_context context)
}
+/* OpenMP 5.0
+ #pragma omp declare mapper ([mapper-identifier :] type var) \
+ [clause [ [,] clause ] ... ] new-line */
+
+static void
+c_parser_omp_declare_mapper (c_parser *parser, enum pragma_context context)
+{
+ tree type, mapper_name = NULL_TREE, var = NULL_TREE, stmt, stmtlist;
+ tree maplist = NULL_TREE, mapper_id, mapper_decl, t;
+ c_token *token;
+
+ if (context == pragma_struct || context == pragma_param)
+ {
+ error ("%<#pragma omp declare reduction%> not at file or block scope");
+ goto fail;
+ }
+
+ if (!c_parser_require (parser, CPP_OPEN_PAREN, "expected %<(%>"))
+ goto fail;
+
+ token = c_parser_peek_token (parser);
+
+ if (c_parser_peek_2nd_token (parser)->type == CPP_COLON)
+ {
+ switch (token->type)
+ {
+ case CPP_NAME:
+ mapper_name = token->value;
+ c_parser_consume_token (parser);
+ break;
+ case CPP_KEYWORD:
+ if (token->keyword == RID_DEFAULT)
+ {
+ mapper_name = NULL_TREE;
+ c_parser_consume_token (parser);
+ break;
+ }
+ /* Fallthrough. */
+ default:
+ error_at (token->location, "expected identifier or %<default%>");
+ c_parser_skip_to_pragma_eol (parser, false);
+ return;
+ }
+
+ if (!c_parser_require (parser, CPP_COLON, "expected %<:%>"))
+ goto fail;
+ }
+
+ mapper_id = c_omp_mapper_id (mapper_name);
+ mapper_decl = c_omp_mapper_decl (mapper_id);
+
+ {
+ location_t loc = c_parser_peek_token (parser)->location;
+ struct c_type_name *ctype = c_parser_type_name (parser);
+ type = groktypename (ctype, NULL, NULL);
+ if (type == error_mark_node)
+ goto fail;
+ if (TREE_CODE (type) != RECORD_TYPE
+ && TREE_CODE (type) != UNION_TYPE)
+ {
+ error_at (loc, "%qT is not a struct or union type in "
+ "%<#pragma omp declare mapper%>", type);
+ c_parser_skip_to_pragma_eol (parser, false);
+ return;
+ }
+ for (tree t = DECL_INITIAL (mapper_decl); t; t = TREE_CHAIN (t))
+ if (comptypes (TREE_PURPOSE (t), type))
+ {
+ error_at (loc, "redeclaration of %qs %<#pragma omp declare "
+ "mapper%> for type %qT", IDENTIFIER_POINTER (mapper_id)
+ + sizeof ("omp declare mapper ") - 1,
+ type);
+ tree prevmapper = TREE_VALUE (t);
+ /* Hmm, this location might not be very accurate. */
+ location_t ploc
+ = DECL_SOURCE_LOCATION (OMP_DECLARE_MAPPER_DECL (prevmapper));
+ error_at (ploc, "previous %<#pragma omp declare mapper%>");
+ c_parser_skip_to_pragma_eol (parser, false);
+ return;
+ }
+ }
+
+ token = c_parser_peek_token (parser);
+ if (token->type == CPP_NAME)
+ {
+ var = build_decl (token->location, VAR_DECL, token->value, type);
+ c_parser_consume_token (parser);
+ DECL_ARTIFICIAL (var) = 1;
+ }
+ else
+ {
+ error_at (token->location, "expected identifier");
+ goto fail;
+ }
+
+ if (!c_parser_require (parser, CPP_CLOSE_PAREN, "expected %<)%>"))
+ goto fail;
+
+ push_scope ();
+ stmtlist = push_stmt_list ();
+ pushdecl (var);
+ DECL_CONTEXT (var) = current_function_decl;
+
+ while (c_parser_next_token_is_not (parser, CPP_PRAGMA_EOL))
+ {
+ location_t here;
+ pragma_omp_clause c_kind;
+ here = c_parser_peek_token (parser)->location;
+ c_kind = c_parser_omp_clause_name (parser);
+ if (c_kind != PRAGMA_OMP_CLAUSE_MAP)
+ {
+ error_at (here, "unexpected clause");
+ goto fail;
+ }
+ maplist = c_parser_omp_clause_map (parser, maplist, GOMP_MAP_UNSET);
+ }
+
+ if (maplist == NULL_TREE)
+ {
+ error_at (input_location, "missing %<map%> clause");
+ goto fail;
+ }
+
+ stmt = make_node (OMP_DECLARE_MAPPER);
+ TREE_TYPE (stmt) = type;
+ OMP_DECLARE_MAPPER_ID (stmt) = mapper_name;
+ OMP_DECLARE_MAPPER_DECL (stmt) = var;
+ OMP_DECLARE_MAPPER_CLAUSES (stmt) = maplist;
+
+ add_stmt (stmt);
+
+ pop_stmt_list (stmtlist);
+ pop_scope ();
+
+ c_parser_skip_to_pragma_eol (parser);
+
+ t = tree_cons (type, stmt, DECL_INITIAL (mapper_decl));
+ DECL_INITIAL (mapper_decl) = t;
+
+ return;
+
+ fail:
+ c_parser_skip_to_pragma_eol (parser);
+}
+
/* OpenMP 4.0
#pragma omp declare simd declare-simd-clauses[optseq] new-line
#pragma omp declare reduction (reduction-id : typename-list : expression) \
@@ -23673,6 +23912,12 @@ c_parser_omp_declare (c_parser *parser, enum pragma_context context)
c_parser_omp_declare_reduction (parser, context);
return false;
}
+ if (strcmp (p, "mapper") == 0)
+ {
+ c_parser_consume_token (parser);
+ c_parser_omp_declare_mapper (parser, context);
+ return false;
+ }
if (!flag_openmp) /* flag_openmp_simd */
{
c_parser_skip_to_pragma_eol (parser, false);
@@ -813,6 +813,10 @@ extern tree c_finish_omp_task (location_t, tree, tree);
extern void c_finish_omp_cancel (location_t, tree);
extern void c_finish_omp_cancellation_point (location_t, tree);
extern tree c_finish_omp_clauses (tree, enum c_omp_region_type);
+extern tree c_omp_finish_mapper_clauses (tree);
+extern tree c_omp_mapper_lookup (tree, tree);
+extern tree c_omp_extract_mapper_directive (tree);
+extern tree c_omp_map_array_section (location_t, tree);
extern tree c_build_va_arg (location_t, tree, location_t, tree);
extern tree c_finish_transaction (location_t, tree, int);
extern bool c_tree_equal (tree, tree);
@@ -864,6 +868,10 @@ extern tree c_omp_reduction_id (enum tree_code, tree);
extern tree c_omp_reduction_decl (tree);
extern tree c_omp_reduction_lookup (tree, tree);
extern tree c_check_omp_declare_reduction_r (tree *, int *, void *);
+extern tree c_omp_mapper_id (tree);
+extern tree c_omp_mapper_decl (tree);
+extern void c_omp_scan_mapper_bindings (location_t, tree *, tree);
+extern tree c_omp_instantiate_mappers (tree);
extern bool c_check_in_current_scope (tree);
extern void c_pushtag (location_t, tree, tree);
extern void c_bind (location_t, tree, bool);
@@ -15336,6 +15336,12 @@ c_finish_omp_clauses (tree clauses, enum c_omp_region_type ort)
case OMP_CLAUSE_MAP:
if (OMP_CLAUSE_MAP_IMPLICIT (c) && !implicit_moved)
goto move_implicit;
+ if (OMP_CLAUSE_MAP_KIND (c) == GOMP_MAP_PUSH_MAPPER_NAME
+ || OMP_CLAUSE_MAP_KIND (c) == GOMP_MAP_POP_MAPPER_NAME)
+ {
+ remove = true;
+ break;
+ }
/* FALLTHRU */
case OMP_CLAUSE_TO:
case OMP_CLAUSE_FROM:
@@ -16178,6 +16184,15 @@ c_finish_omp_clauses (tree clauses, enum c_omp_region_type ort)
return clauses;
}
+/* Do processing necessary to make CLAUSES well-formed, where CLAUSES result
+ from implicit instantiation of user-defined mappers (in gimplify.cc). */
+
+tree
+c_omp_finish_mapper_clauses (tree clauses)
+{
+ return c_finish_omp_clauses (clauses, C_ORT_OMP);
+}
+
/* Return code to initialize DST with a copy constructor from SRC.
C doesn't have copy constructors nor assignment operators, only for
_Atomic vars we need to perform __atomic_load from src into a temporary
@@ -1,4 +1,4 @@
-/* { dg-do compile { target c++ } } */
+/* { dg-do compile } */
struct XYZ {
int a;
@@ -1,4 +1,4 @@
-/* { dg-do compile { target c++ } } */
+/* { dg-do compile } */
/* { dg-additional-options "-fdump-tree-original" } */
/* Check mapper binding clauses. */
@@ -1,4 +1,4 @@
-/* { dg-do compile { target c++ } } */
+/* { dg-do compile } */
typedef struct S_ {
int *myarr;
@@ -1,4 +1,4 @@
-/* { dg-do compile { target c++ } } */
+/* { dg-do compile } */
int x = 5;
@@ -1,4 +1,4 @@
-/* { dg-do compile { target c++ } } */
+/* { dg-do compile } */
struct Q {
int *arr1;
@@ -1,4 +1,4 @@
-/* { dg-do compile { target c++ } } */
+/* { dg-do compile } */
struct Q {
int *arr1;
@@ -1,4 +1,4 @@
-/* { dg-do compile { target c++ } } */
+/* { dg-do compile } */
int x = 5;
new file mode 100644
@@ -0,0 +1,61 @@
+/* { dg-do compile } */
+/* { dg-additional-options "-fdump-tree-gimple" } */
+
+// "omp declare mapper" support -- check expansion in gimple.
+
+#include <stdlib.h>
+
+struct S {
+ int *ptr;
+ int size;
+};
+
+#define N 64
+
+#pragma omp declare mapper (struct S w) map(w.size, w.ptr, w.ptr[:w.size])
+#pragma omp declare mapper (foo:struct S w) map(to:w.size, w.ptr) \
+ map(w.ptr[:w.size])
+
+int main (int argc, char *argv[])
+{
+ struct S s;
+ s.ptr = (int *) malloc (sizeof (int) * N);
+ s.size = N;
+
+#pragma omp declare mapper (bar:struct S w) map(w.size, w.ptr, w.ptr[:w.size])
+
+#pragma omp target
+ {
+ for (int i = 0; i < N; i++)
+ s.ptr[i]++;
+ }
+
+#pragma omp target map(tofrom: s)
+ {
+ for (int i = 0; i < N; i++)
+ s.ptr[i]++;
+ }
+
+#pragma omp target map(mapper(default), tofrom: s)
+ {
+ for (int i = 0; i < N; i++)
+ s.ptr[i]++;
+ }
+
+#pragma omp target map(mapper(foo), alloc: s)
+ {
+ for (int i = 0; i < N; i++)
+ s.ptr[i]++;
+ }
+
+#pragma omp target map(mapper(bar), tofrom: s)
+ {
+ for (int i = 0; i < N; i++)
+ s.ptr[i]++;
+ }
+
+ return 0;
+}
+
+/* { dg-final { scan-tree-dump-times {map\(struct:s \[len: 2\]\) map\(tofrom:s\.ptr \[len: [0-9]+\]\) map\(tofrom:s\.size \[len: [0-9]+\]\) map\(tofrom:\*_[0-9]+ \[len: _[0-9]+\]\) map\(attach:s\.ptr \[bias: 0\]\)} 4 "gimple" { target c++ } } } */
+/* { dg-final { scan-tree-dump-times {map\(struct:s \[len: 2\]\) map\(to:s\.ptr \[len: [0-9]+\]\) map\(to:s\.size \[len: [0-9]+\]\) map\(alloc:\*_[0-9]+ \[len: _[0-9]+\]\) map\(attach:s\.ptr \[bias: 0\]\)} 1 "gimple" { target c++ } } } */
new file mode 100644
@@ -0,0 +1,33 @@
+// { dg-do compile }
+
+// Error-checking tests for "omp declare mapper".
+
+typedef struct {
+ int *ptr;
+ int size;
+} S;
+
+typedef struct {
+ int z;
+} Z;
+
+int main (int argc, char *argv[])
+{
+#pragma omp declare mapper (S v) map(v.size, v.ptr[:v.size])
+/* { dg-error "previous '#pragma omp declare mapper'" "" { target c } .-1 } */
+
+ /* This one's a duplicate. */
+#pragma omp declare mapper (default: S v) map (to: v.size) map (v)
+/* { dg-error "redeclaration of '<default>' '#pragma omp declare mapper' for type 'S'" "" { target c } .-1 } */
+
+ /* ...and this one doesn't use a "base language identifier" for the mapper
+ name. */
+#pragma omp declare mapper (case: S v) map (to: v.size)
+/* { dg-error "expected identifier or 'default'" "" { target c } .-1 } */
+
+ /* A non-struct/class/union type isn't supposed to work. */
+#pragma omp declare mapper (name:Z [5]foo) map (foo[0].z)
+/* { dg-error "'Z\\\[5\\\]' is not a struct or union type in '#pragma omp declare mapper'" "" { target c } .-1 } */
+
+ return 0;
+}
@@ -1,4 +1,4 @@
-/* { dg-do run { target c++ } } */
+/* { dg-do run } */
#include <string.h>
#include <stdlib.h>
@@ -1,4 +1,4 @@
-/* { dg-do run { target c++ } } */
+/* { dg-do run } */
#include <string.h>
#include <stdlib.h>
@@ -1,4 +1,4 @@
-/* { dg-do run { target c++ } } */
+/* { dg-do run } */
#include <string.h>
#include <stdlib.h>
@@ -1,4 +1,4 @@
-/* { dg-do run { target c++ } } */
+/* { dg-do run } */
#include <assert.h>
@@ -1,4 +1,4 @@
-/* { dg-do run { target c++ } } */
+/* { dg-do run } */
#include <stdlib.h>
#include <assert.h>
@@ -1,4 +1,4 @@
-/* { dg-do run { target c++ } } */
+/* { dg-do run } */
#include <string.h>
#include <stdlib.h>