Hi!
OpenMP 5.0 will extend the defaultmap clause from always just the
default(tofrom:scalar) form to have various default behaviors before the
: and various categories after the : (plus no : category to cover all
categories).
There are some issues I've raised on omp-lang, this patch implements what I
think is the general intent.
Tested on x86_64-linux, committed to gomp-5_0-branch.
2018-05-03 Jakub Jelinek <ja...@redhat.com>
* tree-core.h (enum omp_clause_defaultmap_kind): New.
(struct tree_omp_clause): Add subcode.defaultmap_kind.
* tree.h (OMP_CLAUSE_DEFAULTMAP_KIND, OMP_CLAUSE_DEFAULTMAP_CATEGORY,
OMP_CLAUSE_DEFAULTMAP_BEHAVIOR, OMP_CLAUSE_DEFAULTMAP_SET_KIND):
Define.
* tree-pretty-print.c (dump_omp_clause): Handle new kinds of
OMP_CLAUSE_DEFAULTMAP.
* gimplify.c (enum gimplify_omp_var_data): Add GOVD_MAP_ALLOC_ONLY
and GOVD_MAP_FROM_ONLY.
(enum gimplify_defaultmap_kind): New.
(struct gimplify_omp_ctx): Remove target_map_scalars_firstprivate and
target_map_pointers_as_0len_arrays members, add defaultmap.
(new_omp_context): Initialize defaultmap member.
(omp_firstprivatize_variable): Test ctx->defaultmap[GDMK_SCALAR]
instead of ctx->omp_firstprivatize_variable.
(computable_teams_clause): Likewise.
(omp_notice_variable): Handle new defaultmap clause kinds.
(gimplify_scan_omp_clauses): Likewise.
(gimplify_adjust_omp_clauses_1): Handle GOVD_MAP_ALLOC_ONLY
and GOVD_MAP_FROM_ONLY.
c/
* c-parser.c (c_parser_omp_clause_defaultmap): Parse new kinds of
defaultmap clause.
cp/
* parser.c (cp_parser_omp_clause_defaultmap): Parse new kinds of
defaultmap clause.
fortran/
* trans-openmp.c (gfc_trans_omp_clauses): Use
OMP_CLAUSE_DEFAULTMAP_SET_KIND.
testsuite/
* c-c++-common/gomp/defaultmap-1.c: New test.
* c-c++-common/gomp/defaultmap-2.c: New test.
* c-c++-common/gomp/defaultmap-3.c: New test.
--- gcc/tree-core.h.jj 2018-04-30 14:21:18.317028050 +0200
+++ gcc/tree-core.h 2018-05-02 17:29:55.902260817 +0200
@@ -507,6 +507,25 @@ enum omp_clause_default_kind {
OMP_CLAUSE_DEFAULT_LAST
};
+enum omp_clause_defaultmap_kind {
+ OMP_CLAUSE_DEFAULTMAP_CATEGORY_UNSPECIFIED,
+ OMP_CLAUSE_DEFAULTMAP_CATEGORY_SCALAR,
+ OMP_CLAUSE_DEFAULTMAP_CATEGORY_AGGREGATE,
+ OMP_CLAUSE_DEFAULTMAP_CATEGORY_ALLOCATABLE,
+ OMP_CLAUSE_DEFAULTMAP_CATEGORY_POINTER,
+ OMP_CLAUSE_DEFAULTMAP_CATEGORY_MASK = 7,
+ OMP_CLAUSE_DEFAULTMAP_ALLOC = 1 * (OMP_CLAUSE_DEFAULTMAP_CATEGORY_MASK + 1),
+ OMP_CLAUSE_DEFAULTMAP_TO = 2 * (OMP_CLAUSE_DEFAULTMAP_CATEGORY_MASK + 1),
+ OMP_CLAUSE_DEFAULTMAP_FROM = 3 * (OMP_CLAUSE_DEFAULTMAP_CATEGORY_MASK + 1),
+ OMP_CLAUSE_DEFAULTMAP_TOFROM = 4 * (OMP_CLAUSE_DEFAULTMAP_CATEGORY_MASK + 1),
+ OMP_CLAUSE_DEFAULTMAP_FIRSTPRIVATE
+ = 5 * (OMP_CLAUSE_DEFAULTMAP_CATEGORY_MASK + 1),
+ OMP_CLAUSE_DEFAULTMAP_NONE = 6 * (OMP_CLAUSE_DEFAULTMAP_CATEGORY_MASK + 1),
+ OMP_CLAUSE_DEFAULTMAP_DEFAULT
+ = 7 * (OMP_CLAUSE_DEFAULTMAP_CATEGORY_MASK + 1),
+ OMP_CLAUSE_DEFAULTMAP_MASK = 7 * (OMP_CLAUSE_DEFAULTMAP_CATEGORY_MASK + 1)
+};
+
/* There is a TYPE_QUAL value for each type qualifier. They can be
combined by bitwise-or to form the complete set of qualifiers for a
type. */
@@ -1476,6 +1495,7 @@ struct GTY(()) tree_omp_clause {
enum tree_code reduction_code;
enum omp_clause_linear_kind linear_kind;
enum tree_code if_modifier;
+ enum omp_clause_defaultmap_kind defaultmap_kind;
/* The dimension a OMP_CLAUSE__GRIDDIM_ clause of a gridified target
construct describes. */
unsigned int dimension;
--- gcc/tree.h.jj 2018-04-30 13:49:45.209825769 +0200
+++ gcc/tree.h 2018-05-02 17:51:03.371341293 +0200
@@ -1695,6 +1695,18 @@ extern tree maybe_wrap_with_location (tr
#define OMP_CLAUSE_DEFAULT_KIND(NODE) \
(OMP_CLAUSE_SUBCODE_CHECK (NODE,
OMP_CLAUSE_DEFAULT)->omp_clause.subcode.default_kind)
+#define OMP_CLAUSE_DEFAULTMAP_KIND(NODE) \
+ (OMP_CLAUSE_SUBCODE_CHECK (NODE,
OMP_CLAUSE_DEFAULTMAP)->omp_clause.subcode.defaultmap_kind)
+#define OMP_CLAUSE_DEFAULTMAP_CATEGORY(NODE) \
+ ((enum omp_clause_defaultmap_kind) \
+ (OMP_CLAUSE_DEFAULTMAP_KIND (NODE) & OMP_CLAUSE_DEFAULTMAP_CATEGORY_MASK))
+#define OMP_CLAUSE_DEFAULTMAP_BEHAVIOR(NODE) \
+ ((enum omp_clause_defaultmap_kind) \
+ (OMP_CLAUSE_DEFAULTMAP_KIND (NODE) & OMP_CLAUSE_DEFAULTMAP_MASK))
+#define OMP_CLAUSE_DEFAULTMAP_SET_KIND(NODE, BEHAVIOR, CATEGORY) \
+ (OMP_CLAUSE_DEFAULTMAP_KIND (NODE) \
+ = (enum omp_clause_defaultmap_kind) (CATEGORY | BEHAVIOR))
+
#define OMP_CLAUSE_TILE_LIST(NODE) \
OMP_CLAUSE_OPERAND (OMP_CLAUSE_SUBCODE_CHECK (NODE, OMP_CLAUSE_TILE), 0)
#define OMP_CLAUSE_TILE_ITERVAR(NODE) \
--- gcc/tree-pretty-print.c.jj 2018-04-30 13:49:23.109778102 +0200
+++ gcc/tree-pretty-print.c 2018-05-02 17:35:03.111530014 +0200
@@ -901,7 +901,53 @@ dump_omp_clause (pretty_printer *pp, tre
break;
case OMP_CLAUSE_DEFAULTMAP:
- pp_string (pp, "defaultmap(tofrom:scalar)");
+ pp_string (pp, "defaultmap(");
+ switch (OMP_CLAUSE_DEFAULTMAP_BEHAVIOR (clause))
+ {
+ case OMP_CLAUSE_DEFAULTMAP_ALLOC:
+ pp_string (pp, "alloc");
+ break;
+ case OMP_CLAUSE_DEFAULTMAP_TO:
+ pp_string (pp, "to");
+ break;
+ case OMP_CLAUSE_DEFAULTMAP_FROM:
+ pp_string (pp, "from");
+ break;
+ case OMP_CLAUSE_DEFAULTMAP_TOFROM:
+ pp_string (pp, "tofrom");
+ break;
+ case OMP_CLAUSE_DEFAULTMAP_FIRSTPRIVATE:
+ pp_string (pp, "firstprivate");
+ break;
+ case OMP_CLAUSE_DEFAULTMAP_NONE:
+ pp_string (pp, "none");
+ break;
+ case OMP_CLAUSE_DEFAULTMAP_DEFAULT:
+ pp_string (pp, "default");
+ break;
+ default:
+ gcc_unreachable ();
+ }
+ switch (OMP_CLAUSE_DEFAULTMAP_CATEGORY (clause))
+ {
+ case OMP_CLAUSE_DEFAULTMAP_CATEGORY_UNSPECIFIED:
+ break;
+ case OMP_CLAUSE_DEFAULTMAP_CATEGORY_SCALAR:
+ pp_string (pp, ":scalar");
+ break;
+ case OMP_CLAUSE_DEFAULTMAP_CATEGORY_AGGREGATE:
+ pp_string (pp, ":aggregate");
+ break;
+ case OMP_CLAUSE_DEFAULTMAP_CATEGORY_ALLOCATABLE:
+ pp_string (pp, ":allocatable");
+ break;
+ case OMP_CLAUSE_DEFAULTMAP_CATEGORY_POINTER:
+ pp_string (pp, ":pointer");
+ break;
+ default:
+ gcc_unreachable ();
+ }
+ pp_right_paren (pp);
break;
case OMP_CLAUSE__SIMDUID_:
--- gcc/gimplify.c.jj 2018-04-30 14:20:12.673001091 +0200
+++ gcc/gimplify.c 2018-05-03 15:22:54.521416751 +0200
@@ -105,6 +105,12 @@ enum gimplify_omp_var_data
/* Flag for GOVD_MAP: must be present already. */
GOVD_MAP_FORCE_PRESENT = 524288,
+ /* Flag for GOVD_MAP: only allocate. */
+ GOVD_MAP_ALLOC_ONLY = 1048576,
+
+ /* Flag for GOVD_MAP: only copy back. */
+ GOVD_MAP_FROM_ONLY = 2097152,
+
GOVD_DATA_SHARE_CLASS = (GOVD_SHARED | GOVD_PRIVATE | GOVD_FIRSTPRIVATE
| GOVD_LASTPRIVATE | GOVD_REDUCTION | GOVD_LINEAR
| GOVD_LOCAL)
@@ -176,6 +182,14 @@ struct gimplify_ctx
unsigned in_switch_expr : 1;
};
+enum gimplify_defaultmap_kind
+{
+ GDMK_SCALAR,
+ GDMK_AGGREGATE,
+ GDMK_ALLOCATABLE,
+ GDMK_POINTER
+};
+
struct gimplify_omp_ctx
{
struct gimplify_omp_ctx *outer_context;
@@ -188,9 +202,8 @@ struct gimplify_omp_ctx
enum omp_region_type region_type;
bool combined_loop;
bool distribute;
- bool target_map_scalars_firstprivate;
- bool target_map_pointers_as_0len_arrays;
bool target_firstprivatize_array_bases;
+ int defaultmap[4];
};
static struct gimplify_ctx *gimplify_ctxp;
@@ -413,6 +426,10 @@ new_omp_context (enum omp_region_type re
c->default_kind = OMP_CLAUSE_DEFAULT_SHARED;
else
c->default_kind = OMP_CLAUSE_DEFAULT_UNSPECIFIED;
+ c->defaultmap[GDMK_SCALAR] = GOVD_MAP;
+ c->defaultmap[GDMK_AGGREGATE] = GOVD_MAP;
+ c->defaultmap[GDMK_ALLOCATABLE] = GOVD_MAP;
+ c->defaultmap[GDMK_POINTER] = GOVD_MAP;
return c;
}
@@ -6685,7 +6702,7 @@ omp_firstprivatize_variable (struct gimp
}
else if ((ctx->region_type & ORT_TARGET) != 0)
{
- if (ctx->target_map_scalars_firstprivate)
+ if (ctx->defaultmap[GDMK_SCALAR] & GOVD_FIRSTPRIVATE)
omp_add_variable (ctx, decl, GOVD_FIRSTPRIVATE);
else
omp_add_variable (ctx, decl, GOVD_MAP | GOVD_MAP_TO_ONLY);
@@ -7217,11 +7234,9 @@ omp_notice_variable (struct gimplify_omp
if (n == NULL)
{
unsigned nflags = flags;
- if (ctx->target_map_pointers_as_0len_arrays
- || ctx->target_map_scalars_firstprivate)
+ if ((ctx->region_type & ORT_ACC) == 0)
{
bool is_declare_target = false;
- bool is_scalar = false;
if (is_global_var (decl)
&& varpool_node::get_create (decl)->offloadable)
{
@@ -7238,18 +7253,34 @@ omp_notice_variable (struct gimplify_omp
}
is_declare_target = octx == NULL;
}
- if (!is_declare_target && ctx->target_map_scalars_firstprivate)
- is_scalar = lang_hooks.decls.omp_scalar_p (decl);
- if (is_declare_target)
- ;
- else if (ctx->target_map_pointers_as_0len_arrays
- && (TREE_CODE (TREE_TYPE (decl)) == POINTER_TYPE
- || (TREE_CODE (TREE_TYPE (decl)) == REFERENCE_TYPE
- && TREE_CODE (TREE_TYPE (TREE_TYPE (decl)))
- == POINTER_TYPE)))
- nflags |= GOVD_MAP | GOVD_MAP_0LEN_ARRAY;
- else if (is_scalar)
- nflags |= GOVD_FIRSTPRIVATE;
+ if (!is_declare_target)
+ {
+ int gdmk;
+ if (TREE_CODE (TREE_TYPE (decl)) == POINTER_TYPE
+ || (TREE_CODE (TREE_TYPE (decl)) == REFERENCE_TYPE
+ && (TREE_CODE (TREE_TYPE (TREE_TYPE (decl)))
+ == POINTER_TYPE)))
+ gdmk = GDMK_POINTER;
+ else if (lang_hooks.decls.omp_scalar_p (decl))
+ gdmk = GDMK_SCALAR;
+ else
+ gdmk = GDMK_AGGREGATE;
+ if (ctx->defaultmap[gdmk] == 0)
+ {
+ tree d = lang_hooks.decls.omp_report_decl (decl);
+ error ("%qE not specified in enclosing %<target%>",
+ DECL_NAME (d));
+ error_at (ctx->location, "enclosing %<target%>");
+ }
+ else if (ctx->defaultmap[gdmk]
+ & (GOVD_MAP_0LEN_ARRAY | GOVD_FIRSTPRIVATE))
+ nflags |= ctx->defaultmap[gdmk];
+ else
+ {
+ gcc_assert (ctx->defaultmap[gdmk] & GOVD_MAP);
+ nflags |= ctx->defaultmap[gdmk] & ~GOVD_MAP;
+ }
+ }
}
struct gimplify_omp_ctx *octx = ctx->outer_context;
@@ -7280,28 +7311,28 @@ omp_notice_variable (struct gimplify_omp
}
}
- {
- tree type = TREE_TYPE (decl);
+ if ((nflags & ~(GOVD_MAP_TO_ONLY | GOVD_MAP_FROM_ONLY
+ | GOVD_MAP_ALLOC_ONLY)) == flags)
+ {
+ tree type = TREE_TYPE (decl);
- if (nflags == flags
- && gimplify_omp_ctxp->target_firstprivatize_array_bases
- && lang_hooks.decls.omp_privatize_by_reference (decl))
- type = TREE_TYPE (type);
- if (nflags == flags
- && !lang_hooks.types.omp_mappable_type (type))
- {
- error ("%qD referenced in target region does not have "
- "a mappable type", decl);
- nflags |= GOVD_MAP | GOVD_EXPLICIT;
- }
- else if (nflags == flags)
- {
- if ((ctx->region_type & ORT_ACC) != 0)
- nflags = oacc_default_clause (ctx, decl, flags);
- else
- nflags |= GOVD_MAP;
- }
- }
+ if (gimplify_omp_ctxp->target_firstprivatize_array_bases
+ && lang_hooks.decls.omp_privatize_by_reference (decl))
+ type = TREE_TYPE (type);
+ if (!lang_hooks.types.omp_mappable_type (type))
+ {
+ error ("%qD referenced in target region does not have "
+ "a mappable type", decl);
+ nflags |= GOVD_MAP | GOVD_EXPLICIT;
+ }
+ else
+ {
+ if ((ctx->region_type & ORT_ACC) != 0)
+ nflags = oacc_default_clause (ctx, decl, flags);
+ else
+ nflags |= GOVD_MAP;
+ }
+ }
found_outer:
omp_add_variable (ctx, decl, nflags);
}
@@ -7545,8 +7576,8 @@ gimplify_scan_omp_clauses (tree *list_p,
if (code == OMP_TARGET)
{
if (!lang_GNU_Fortran ())
- ctx->target_map_pointers_as_0len_arrays = true;
- ctx->target_map_scalars_firstprivate = true;
+ ctx->defaultmap[GDMK_POINTER] = GOVD_MAP | GOVD_MAP_0LEN_ARRAY;
+ ctx->defaultmap[GDMK_SCALAR] = GOVD_FIRSTPRIVATE;
}
if (!lang_GNU_Fortran ())
switch (code)
@@ -8599,7 +8630,69 @@ gimplify_scan_omp_clauses (tree *list_p,
break;
case OMP_CLAUSE_DEFAULTMAP:
- ctx->target_map_scalars_firstprivate = false;
+ enum gimplify_defaultmap_kind gdmkmin, gdmkmax;
+ switch (OMP_CLAUSE_DEFAULTMAP_CATEGORY (c))
+ {
+ case OMP_CLAUSE_DEFAULTMAP_CATEGORY_UNSPECIFIED:
+ gdmkmin = GDMK_SCALAR;
+ gdmkmax = GDMK_POINTER;
+ break;
+ case OMP_CLAUSE_DEFAULTMAP_CATEGORY_SCALAR:
+ gdmkmin = gdmkmax = GDMK_SCALAR;
+ break;
+ case OMP_CLAUSE_DEFAULTMAP_CATEGORY_AGGREGATE:
+ gdmkmin = gdmkmax = GDMK_AGGREGATE;
+ break;
+ case OMP_CLAUSE_DEFAULTMAP_CATEGORY_ALLOCATABLE:
+ gdmkmin = gdmkmax = GDMK_ALLOCATABLE;
+ break;
+ case OMP_CLAUSE_DEFAULTMAP_CATEGORY_POINTER:
+ gdmkmin = gdmkmax = GDMK_POINTER;
+ break;
+ default:
+ gcc_unreachable ();
+ }
+ for (int gdmk = gdmkmin; gdmk <= gdmkmax; gdmk++)
+ switch (OMP_CLAUSE_DEFAULTMAP_BEHAVIOR (c))
+ {
+ case OMP_CLAUSE_DEFAULTMAP_ALLOC:
+ ctx->defaultmap[gdmk] = GOVD_MAP | GOVD_MAP_ALLOC_ONLY;
+ break;
+ case OMP_CLAUSE_DEFAULTMAP_TO:
+ ctx->defaultmap[gdmk] = GOVD_MAP | GOVD_MAP_TO_ONLY;
+ break;
+ case OMP_CLAUSE_DEFAULTMAP_FROM:
+ ctx->defaultmap[gdmk] = GOVD_MAP | GOVD_MAP_FROM_ONLY;
+ break;
+ case OMP_CLAUSE_DEFAULTMAP_TOFROM:
+ ctx->defaultmap[gdmk] = GOVD_MAP;
+ break;
+ case OMP_CLAUSE_DEFAULTMAP_FIRSTPRIVATE:
+ ctx->defaultmap[gdmk] = GOVD_FIRSTPRIVATE;
+ break;
+ case OMP_CLAUSE_DEFAULTMAP_NONE:
+ ctx->defaultmap[gdmk] = 0;
+ break;
+ case OMP_CLAUSE_DEFAULTMAP_DEFAULT:
+ switch (gdmk)
+ {
+ case GDMK_SCALAR:
+ ctx->defaultmap[gdmk] = GOVD_FIRSTPRIVATE;
+ break;
+ case GDMK_AGGREGATE:
+ case GDMK_ALLOCATABLE:
+ ctx->defaultmap[gdmk] = GOVD_MAP;
+ break;
+ case GDMK_POINTER:
+ ctx->defaultmap[gdmk] = GOVD_MAP | GOVD_MAP_0LEN_ARRAY;
+ break;
+ default:
+ gcc_unreachable ();
+ }
+ break;
+ default:
+ gcc_unreachable ();
+ }
break;
case OMP_CLAUSE_ALIGNED:
@@ -8898,7 +8991,9 @@ gimplify_adjust_omp_clauses_1 (splay_tre
/* Not all combinations of these GOVD_MAP flags are actually valid. */
switch (flags & (GOVD_MAP_TO_ONLY
| GOVD_MAP_FORCE
- | GOVD_MAP_FORCE_PRESENT))
+ | GOVD_MAP_FORCE_PRESENT
+ | GOVD_MAP_ALLOC_ONLY
+ | GOVD_MAP_FROM_ONLY))
{
case 0:
kind = GOMP_MAP_TOFROM;
@@ -8909,6 +9004,12 @@ gimplify_adjust_omp_clauses_1 (splay_tre
case GOVD_MAP_TO_ONLY:
kind = GOMP_MAP_TO;
break;
+ case GOVD_MAP_FROM_ONLY:
+ kind = GOMP_MAP_FROM;
+ break;
+ case GOVD_MAP_ALLOC_ONLY:
+ kind = GOMP_MAP_ALLOC;
+ break;
case GOVD_MAP_TO_ONLY | GOVD_MAP_FORCE:
kind = GOMP_MAP_TO | GOMP_MAP_FLAG_FORCE;
break;
@@ -10589,7 +10690,7 @@ computable_teams_clause (tree *tp, int *
(splay_tree_key) *tp);
if (n == NULL)
{
- if (gimplify_omp_ctxp->target_map_scalars_firstprivate)
+ if (gimplify_omp_ctxp->defaultmap[GDMK_SCALAR] & GOVD_FIRSTPRIVATE)
return NULL_TREE;
return *tp;
}
--- gcc/c/c-parser.c.jj 2018-04-30 15:07:52.972514234 +0200
+++ gcc/c/c-parser.c 2018-05-03 17:16:15.668780501 +0200
@@ -12463,7 +12463,10 @@ c_parser_omp_clause_hint (c_parser *pars
}
/* OpenMP 4.5:
- defaultmap ( tofrom : scalar ) */
+ defaultmap ( tofrom : scalar )
+
+ OpenMP 5.0:
+ defaultmap ( implicit-behavior [ : variable-category ] ) */
static tree
c_parser_omp_clause_defaultmap (c_parser *parser, tree list)
@@ -12471,39 +12474,155 @@ c_parser_omp_clause_defaultmap (c_parser
location_t loc = c_parser_peek_token (parser)->location;
tree c;
const char *p;
+ enum omp_clause_defaultmap_kind behavior = OMP_CLAUSE_DEFAULTMAP_DEFAULT;
+ enum omp_clause_defaultmap_kind category
+ = OMP_CLAUSE_DEFAULTMAP_CATEGORY_UNSPECIFIED;
matching_parens parens;
if (!parens.require_open (parser))
return list;
- if (!c_parser_next_token_is (parser, CPP_NAME))
+ if (c_parser_next_token_is_keyword (parser, RID_DEFAULT))
+ p = "default";
+ else if (!c_parser_next_token_is (parser, CPP_NAME))
{
- c_parser_error (parser, "expected %<tofrom%>");
+ invalid_behavior:
+ c_parser_error (parser, "expected %<alloc%>, %<to%>, %<from%>, "
+ "%<tofrom%>, %<firstprivate%>, %<none%> "
+ "or %<default%>");
goto out_err;
}
- p = IDENTIFIER_POINTER (c_parser_peek_token (parser)->value);
- if (strcmp (p, "tofrom") != 0)
+ else
+ p = IDENTIFIER_POINTER (c_parser_peek_token (parser)->value);
+
+ switch (p[0])
{
- c_parser_error (parser, "expected %<tofrom%>");
- goto out_err;
+ case 'a':
+ if (strcmp ("alloc", p) == 0)
+ behavior = OMP_CLAUSE_DEFAULTMAP_ALLOC;
+ else
+ goto invalid_behavior;
+ break;
+
+ case 'd':
+ if (strcmp ("default", p) == 0)
+ behavior = OMP_CLAUSE_DEFAULTMAP_DEFAULT;
+ else
+ goto invalid_behavior;
+ break;
+
+ case 'f':
+ if (strcmp ("firstprivate", p) == 0)
+ behavior = OMP_CLAUSE_DEFAULTMAP_FIRSTPRIVATE;
+ else if (strcmp ("from", p) == 0)
+ behavior = OMP_CLAUSE_DEFAULTMAP_FROM;
+ else
+ goto invalid_behavior;
+ break;
+
+ case 'n':
+ if (strcmp ("none", p) == 0)
+ behavior = OMP_CLAUSE_DEFAULTMAP_NONE;
+ else
+ goto invalid_behavior;
+ break;
+
+ case 't':
+ if (strcmp ("tofrom", p) == 0)
+ behavior = OMP_CLAUSE_DEFAULTMAP_TOFROM;
+ else if (strcmp ("to", p) == 0)
+ behavior = OMP_CLAUSE_DEFAULTMAP_TO;
+ else
+ goto invalid_behavior;
+ break;
+
+ default:
+ goto invalid_behavior;
}
c_parser_consume_token (parser);
- if (!c_parser_require (parser, CPP_COLON, "expected %<:%>"))
- goto out_err;
- if (!c_parser_next_token_is (parser, CPP_NAME))
- {
- c_parser_error (parser, "expected %<scalar%>");
- goto out_err;
- }
- p = IDENTIFIER_POINTER (c_parser_peek_token (parser)->value);
- if (strcmp (p, "scalar") != 0)
+
+ if (!c_parser_next_token_is (parser, CPP_CLOSE_PAREN))
{
- c_parser_error (parser, "expected %<scalar%>");
- goto out_err;
+ if (!c_parser_require (parser, CPP_COLON, "expected %<:%>"))
+ goto out_err;
+ if (!c_parser_next_token_is (parser, CPP_NAME))
+ {
+ invalid_category:
+ c_parser_error (parser, "expected %<scalar%>, %<aggregate%> or "
+ "%<pointer%>");
+ goto out_err;
+ }
+ p = IDENTIFIER_POINTER (c_parser_peek_token (parser)->value);
+ switch (p[0])
+ {
+ case 'a':
+ if (strcmp ("aggregate", p) == 0)
+ category = OMP_CLAUSE_DEFAULTMAP_CATEGORY_AGGREGATE;
+ else
+ goto invalid_category;
+ break;
+
+ case 'p':
+ if (strcmp ("pointer", p) == 0)
+ category = OMP_CLAUSE_DEFAULTMAP_CATEGORY_POINTER;
+ else
+ goto invalid_category;
+ break;
+
+ case 's':
+ if (strcmp ("scalar", p) == 0)
+ category = OMP_CLAUSE_DEFAULTMAP_CATEGORY_SCALAR;
+ else
+ goto invalid_category;
+ break;
+
+ default:
+ goto invalid_category;
+ }
+
+ c_parser_consume_token (parser);
}
- c_parser_consume_token (parser);
parens.skip_until_found_close (parser);
- check_no_duplicate_clause (list, OMP_CLAUSE_DEFAULTMAP, "defaultmap");
+
+ for (c = list; c ; c = OMP_CLAUSE_CHAIN (c))
+ if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_DEFAULTMAP
+ && (category == OMP_CLAUSE_DEFAULTMAP_CATEGORY_UNSPECIFIED
+ || OMP_CLAUSE_DEFAULTMAP_CATEGORY (c) == category
+ || (OMP_CLAUSE_DEFAULTMAP_CATEGORY (c)
+ == OMP_CLAUSE_DEFAULTMAP_CATEGORY_UNSPECIFIED)))
+ {
+ enum omp_clause_defaultmap_kind cat = category;
+ location_t loc = OMP_CLAUSE_LOCATION (c);
+ if (cat == OMP_CLAUSE_DEFAULTMAP_CATEGORY_UNSPECIFIED)
+ cat = OMP_CLAUSE_DEFAULTMAP_CATEGORY (c);
+ p = NULL;
+ switch (cat)
+ {
+ case OMP_CLAUSE_DEFAULTMAP_CATEGORY_UNSPECIFIED:
+ p = NULL;
+ break;
+ case OMP_CLAUSE_DEFAULTMAP_CATEGORY_AGGREGATE:
+ p = "aggregate";
+ break;
+ case OMP_CLAUSE_DEFAULTMAP_CATEGORY_POINTER:
+ p = "pointer";
+ break;
+ case OMP_CLAUSE_DEFAULTMAP_CATEGORY_SCALAR:
+ p = "scalar";
+ break;
+ default:
+ gcc_unreachable ();
+ }
+ if (p)
+ error_at (loc, "too many %<defaultmap%> clauses with %qs category",
+ p);
+ else
+ error_at (loc, "too many %<defaultmap%> clauses with unspecified "
+ "category");
+ break;
+ }
+
c = build_omp_clause (loc, OMP_CLAUSE_DEFAULTMAP);
+ OMP_CLAUSE_DEFAULTMAP_SET_KIND (c, behavior, category);
OMP_CLAUSE_CHAIN (c) = list;
return c;
--- gcc/cp/parser.c.jj 2018-04-30 16:11:55.438350128 +0200
+++ gcc/cp/parser.c 2018-05-03 17:18:14.432877678 +0200
@@ -32567,7 +32567,10 @@ cp_parser_omp_clause_hint (cp_parser *pa
}
/* OpenMP 4.5:
- defaultmap ( tofrom : scalar ) */
+ defaultmap ( tofrom : scalar )
+
+ OpenMP 5.0:
+ defaultmap ( implicit-behavior [ : variable-category ] ) */
static tree
cp_parser_omp_clause_defaultmap (cp_parser *parser, tree list,
@@ -32575,47 +32578,163 @@ cp_parser_omp_clause_defaultmap (cp_pars
{
tree c, id;
const char *p;
+ enum omp_clause_defaultmap_kind behavior = OMP_CLAUSE_DEFAULTMAP_DEFAULT;
+ enum omp_clause_defaultmap_kind category
+ = OMP_CLAUSE_DEFAULTMAP_CATEGORY_UNSPECIFIED;
matching_parens parens;
if (!parens.require_open (parser))
return list;
- if (!cp_lexer_next_token_is (parser->lexer, CPP_NAME))
+ if (cp_lexer_next_token_is_keyword (parser->lexer, RID_DEFAULT))
+ p = "default";
+ else if (!cp_lexer_next_token_is (parser->lexer, CPP_NAME))
{
- cp_parser_error (parser, "expected %<tofrom%>");
+ invalid_behavior:
+ cp_parser_error (parser, "expected %<alloc%>, %<to%>, %<from%>, "
+ "%<tofrom%>, %<firstprivate%>, %<none%> "
+ "or %<default%>");
goto out_err;
}
- id = cp_lexer_peek_token (parser->lexer)->u.value;
- p = IDENTIFIER_POINTER (id);
- if (strcmp (p, "tofrom") != 0)
+ else
{
- cp_parser_error (parser, "expected %<tofrom%>");
- goto out_err;
+ id = cp_lexer_peek_token (parser->lexer)->u.value;
+ p = IDENTIFIER_POINTER (id);
}
- cp_lexer_consume_token (parser->lexer);
- if (!cp_parser_require (parser, CPP_COLON, RT_COLON))
- goto out_err;
- if (!cp_lexer_next_token_is (parser->lexer, CPP_NAME))
+ switch (p[0])
{
- cp_parser_error (parser, "expected %<scalar%>");
- goto out_err;
+ case 'a':
+ if (strcmp ("alloc", p) == 0)
+ behavior = OMP_CLAUSE_DEFAULTMAP_ALLOC;
+ else
+ goto invalid_behavior;
+ break;
+
+ case 'd':
+ if (strcmp ("default", p) == 0)
+ behavior = OMP_CLAUSE_DEFAULTMAP_DEFAULT;
+ else
+ goto invalid_behavior;
+ break;
+
+ case 'f':
+ if (strcmp ("firstprivate", p) == 0)
+ behavior = OMP_CLAUSE_DEFAULTMAP_FIRSTPRIVATE;
+ else if (strcmp ("from", p) == 0)
+ behavior = OMP_CLAUSE_DEFAULTMAP_FROM;
+ else
+ goto invalid_behavior;
+ break;
+
+ case 'n':
+ if (strcmp ("none", p) == 0)
+ behavior = OMP_CLAUSE_DEFAULTMAP_NONE;
+ else
+ goto invalid_behavior;
+ break;
+
+ case 't':
+ if (strcmp ("tofrom", p) == 0)
+ behavior = OMP_CLAUSE_DEFAULTMAP_TOFROM;
+ else if (strcmp ("to", p) == 0)
+ behavior = OMP_CLAUSE_DEFAULTMAP_TO;
+ else
+ goto invalid_behavior;
+ break;
+
+ default:
+ goto invalid_behavior;
}
- id = cp_lexer_peek_token (parser->lexer)->u.value;
- p = IDENTIFIER_POINTER (id);
- if (strcmp (p, "scalar") != 0)
+ cp_lexer_consume_token (parser->lexer);
+
+ if (!cp_lexer_next_token_is (parser->lexer, CPP_CLOSE_PAREN))
{
- cp_parser_error (parser, "expected %<scalar%>");
- goto out_err;
+ if (!cp_parser_require (parser, CPP_COLON, RT_COLON))
+ goto out_err;
+
+ if (!cp_lexer_next_token_is (parser->lexer, CPP_NAME))
+ {
+ invalid_category:
+ cp_parser_error (parser, "expected %<scalar%>, %<aggregate%> or "
+ "%<pointer%>");
+ goto out_err;
+ }
+ id = cp_lexer_peek_token (parser->lexer)->u.value;
+ p = IDENTIFIER_POINTER (id);
+
+ switch (p[0])
+ {
+ case 'a':
+ if (strcmp ("aggregate", p) == 0)
+ category = OMP_CLAUSE_DEFAULTMAP_CATEGORY_AGGREGATE;
+ else
+ goto invalid_category;
+ break;
+
+ case 'p':
+ if (strcmp ("pointer", p) == 0)
+ category = OMP_CLAUSE_DEFAULTMAP_CATEGORY_POINTER;
+ else
+ goto invalid_category;
+ break;
+
+ case 's':
+ if (strcmp ("scalar", p) == 0)
+ category = OMP_CLAUSE_DEFAULTMAP_CATEGORY_SCALAR;
+ else
+ goto invalid_category;
+ break;
+
+ default:
+ goto invalid_category;
+ }
+
+ cp_lexer_consume_token (parser->lexer);
}
- cp_lexer_consume_token (parser->lexer);
if (!parens.require_close (parser))
goto out_err;
- check_no_duplicate_clause (list, OMP_CLAUSE_DEFAULTMAP, "defaultmap",
- location);
+ for (c = list; c ; c = OMP_CLAUSE_CHAIN (c))
+ if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_DEFAULTMAP
+ && (category == OMP_CLAUSE_DEFAULTMAP_CATEGORY_UNSPECIFIED
+ || OMP_CLAUSE_DEFAULTMAP_CATEGORY (c) == category
+ || (OMP_CLAUSE_DEFAULTMAP_CATEGORY (c)
+ == OMP_CLAUSE_DEFAULTMAP_CATEGORY_UNSPECIFIED)))
+ {
+ enum omp_clause_defaultmap_kind cat = category;
+ location_t loc = OMP_CLAUSE_LOCATION (c);
+ if (cat == OMP_CLAUSE_DEFAULTMAP_CATEGORY_UNSPECIFIED)
+ cat = OMP_CLAUSE_DEFAULTMAP_CATEGORY (c);
+ p = NULL;
+ switch (cat)
+ {
+ case OMP_CLAUSE_DEFAULTMAP_CATEGORY_UNSPECIFIED:
+ p = NULL;
+ break;
+ case OMP_CLAUSE_DEFAULTMAP_CATEGORY_AGGREGATE:
+ p = "aggregate";
+ break;
+ case OMP_CLAUSE_DEFAULTMAP_CATEGORY_POINTER:
+ p = "pointer";
+ break;
+ case OMP_CLAUSE_DEFAULTMAP_CATEGORY_SCALAR:
+ p = "scalar";
+ break;
+ default:
+ gcc_unreachable ();
+ }
+ if (p)
+ error_at (loc, "too many %<defaultmap%> clauses with %qs category",
+ p);
+ else
+ error_at (loc, "too many %<defaultmap%> clauses with unspecified "
+ "category");
+ break;
+ }
c = build_omp_clause (location, OMP_CLAUSE_DEFAULTMAP);
+ OMP_CLAUSE_DEFAULTMAP_SET_KIND (c, behavior, category);
OMP_CLAUSE_CHAIN (c) = list;
return c;
--- gcc/fortran/trans-openmp.c.jj 2018-04-30 14:19:46.187990214 +0200
+++ gcc/fortran/trans-openmp.c 2018-05-02 17:52:01.091390796 +0200
@@ -2866,6 +2866,8 @@ gfc_trans_omp_clauses (stmtblock_t *bloc
if (clauses->defaultmap)
{
c = build_omp_clause (where.lb->location, OMP_CLAUSE_DEFAULTMAP);
+ OMP_CLAUSE_DEFAULTMAP_SET_KIND (c, OMP_CLAUSE_DEFAULTMAP_TOFROM,
+ OMP_CLAUSE_DEFAULTMAP_CATEGORY_SCALAR);
omp_clauses = gfc_trans_add_clause (c, omp_clauses);
}
if (clauses->depend_source)
--- gcc/testsuite/c-c++-common/gomp/defaultmap-1.c.jj 2018-05-03
15:55:27.383889617 +0200
+++ gcc/testsuite/c-c++-common/gomp/defaultmap-1.c 2018-05-03
15:56:34.444943107 +0200
@@ -0,0 +1,30 @@
+void
+foo (void)
+{
+ #pragma omp target defaultmap(alloc) defaultmap(alloc) /* { dg-error
"too many 'defaultmap' clauses with unspecified category" } */
+ ;
+ #pragma omp target defaultmap(to) defaultmap(from) /* { dg-error
"too many 'defaultmap' clauses with unspecified category" } */
+ ;
+ #pragma omp target defaultmap(tofrom) defaultmap(firstprivate:scalar)
/* { dg-error "too many 'defaultmap' clauses with 'scalar' category" } */
+ ;
+ #pragma omp target defaultmap(none:aggregate) defaultmap(alloc:scalar)
defaultmap(none:scalar) /* { dg-error "too many 'defaultmap' clauses with
'scalar' category" } */
+ ;
+ #pragma omp target defaultmap(none : pointer) defaultmap ( none ) /* {
dg-error "too many 'defaultmap' clauses with 'pointer' category" } */
+ ;
+ #pragma omp target defaultmap() /* { dg-error
"expected" } */
+ ;
+ #pragma omp target defaultmap(for) /* { dg-error
"expected" } */
+ ;
+ #pragma omp target defaultmap(blah) /* { dg-error
"expected" } */
+ ;
+ #pragma omp target defaultmap(tofrom:) /* { dg-error
"expected" } */
+ ;
+ #pragma omp target defaultmap(tofrom scalar) /* { dg-error
"expected" } */
+ ;
+ #pragma omp target defaultmap(tofrom,scalar) /* { dg-error
"expected" } */
+ ;
+ #pragma omp target defaultmap(default ;) /* { dg-error
"expected" } */
+ ;
+ #pragma omp target defaultmap(default : qux) /* { dg-error
"expected" } */
+ ;
+}
--- gcc/testsuite/c-c++-common/gomp/defaultmap-2.c.jj 2018-05-03
16:49:20.433478764 +0200
+++ gcc/testsuite/c-c++-common/gomp/defaultmap-2.c 2018-05-03
17:28:24.478376850 +0200
@@ -0,0 +1,131 @@
+/* { dg-do compile } */
+/* { dg-additional-options "-fdump-tree-gimple" } */
+
+struct S { int s; };
+void foo (char *);
+void bar (int, char *, struct S, int *);
+#pragma omp declare target to (bar)
+#define N 16
+
+void
+f1 (int sc1, struct S ag1, int *pt1)
+{
+ char ar1[N];
+ foo (ar1);
+ #pragma omp target
+ bar (sc1, ar1, ag1, pt1);
+/* { dg-final { scan-tree-dump "firstprivate\\(sc1\\)" "gimple" } } */
+/* { dg-final { scan-tree-dump "map\\(tofrom:ar1" "gimple" } } */
+/* { dg-final { scan-tree-dump "map\\(tofrom:ag1" "gimple" } } */
+/* { dg-final { scan-tree-dump "map\\(firstprivate:pt1 .pointer assign"
"gimple" } } */
+}
+
+void
+f2 (int sc2, struct S ag2, int *pt2)
+{
+ char ar2[N];
+ foo (ar2);
+ #pragma omp target firstprivate (sc2, ar2, ag2, pt2) defaultmap (none)
+ bar (sc2, ar2, ag2, pt2);
+/* { dg-final { scan-tree-dump "firstprivate\\(sc2\\)" "gimple" } } */
+/* { dg-final { scan-tree-dump "firstprivate\\(ar2\\)" "gimple" } } */
+/* { dg-final { scan-tree-dump "firstprivate\\(ag2\\)" "gimple" } } */
+/* { dg-final { scan-tree-dump "firstprivate\\(pt2\\)" "gimple" } } */
+}
+
+void
+f3 (int sc3, struct S ag3, int *pt3)
+{
+ char ar3[N];
+ foo (ar3);
+ #pragma omp target defaultmap(none:scalar) defaultmap(none:aggregate) \
+ map (sc3, ar3, ag3, pt3) defaultmap(none:pointer)
+ bar (sc3, ar3, ag3, pt3);
+/* { dg-final { scan-tree-dump "map\\(tofrom:sc3" "gimple" } } */
+/* { dg-final { scan-tree-dump "map\\(tofrom:ar3" "gimple" } } */
+/* { dg-final { scan-tree-dump "map\\(tofrom:ag3" "gimple" } } */
+/* { dg-final { scan-tree-dump "map\\(tofrom:pt3" "gimple" } } */
+}
+
+void
+f4 (int sc4, struct S ag4, int *pt4)
+{
+ char ar4[N];
+ foo (ar4);
+ #pragma omp target defaultmap(tofrom:scalar)
+ bar (sc4, ar4, ag4, pt4);
+/* { dg-final { scan-tree-dump "map\\(tofrom:sc4" "gimple" } } */
+/* { dg-final { scan-tree-dump "map\\(tofrom:ar4" "gimple" } } */
+/* { dg-final { scan-tree-dump "map\\(tofrom:ag4" "gimple" } } */
+/* { dg-final { scan-tree-dump "map\\(firstprivate:pt4 .pointer assign"
"gimple" } } */
+}
+
+void
+f5 (int sc5, struct S ag5, int *pt5)
+{
+ char ar5[N];
+ foo (ar5);
+ #pragma omp target defaultmap(to)
+ bar (sc5, ar5, ag5, pt5);
+/* { dg-final { scan-tree-dump "map\\(to:sc5" "gimple" } } */
+/* { dg-final { scan-tree-dump "map\\(to:ar5" "gimple" } } */
+/* { dg-final { scan-tree-dump "map\\(to:ag5" "gimple" } } */
+/* { dg-final { scan-tree-dump "map\\(to:pt5" "gimple" } } */
+}
+
+void
+f6 (int sc6, struct S ag6, int *pt6)
+{
+ char ar6[N];
+ foo (ar6);
+ #pragma omp target defaultmap(firstprivate)
+ bar (sc6, ar6, ag6, pt6);
+/* { dg-final { scan-tree-dump "firstprivate\\(sc6\\)" "gimple" } } */
+/* { dg-final { scan-tree-dump "firstprivate\\(ar6\\)" "gimple" } } */
+/* { dg-final { scan-tree-dump "firstprivate\\(ag6\\)" "gimple" } } */
+/* { dg-final { scan-tree-dump "firstprivate\\(pt6\\)" "gimple" } } */
+}
+
+void
+f7 (int sc7, struct S ag7, int *pt7)
+{
+ char ar7[N];
+ foo (ar7);
+ #pragma omp target defaultmap(alloc: scalar) defaultmap(from: aggregate)
defaultmap(default: pointer)
+ {
+ int *q = &sc7;
+ *q = 6;
+ ag7.s = 5;
+ int i;
+ for (i = 0; i < N; ++i)
+ ar7[i] = 7;
+ bar (sc7, ar7, ag7, pt7);
+ }
+/* { dg-final { scan-tree-dump "map\\(alloc:sc7" "gimple" } } */
+/* { dg-final { scan-tree-dump "map\\(from:ar7" "gimple" } } */
+/* { dg-final { scan-tree-dump "map\\(from:ag7" "gimple" } } */
+/* { dg-final { scan-tree-dump "map\\(firstprivate:pt7 .pointer assign"
"gimple" } } */
+}
+
+void
+f8 (int sc8, struct S ag8, int *pt8)
+{
+ char ar8[N];
+ foo (ar8);
+ #pragma omp target defaultmap(firstprivate:aggregate)
defaultmap(none:scalar) \
+ defaultmap(tofrom:pointer) map(to: sc8)
+ bar (sc8, ar8, ag8, pt8);
+/* { dg-final { scan-tree-dump "map\\(to:sc8" "gimple" } } */
+/* { dg-final { scan-tree-dump "firstprivate\\(ar8\\)" "gimple" } } */
+/* { dg-final { scan-tree-dump "firstprivate\\(ag8\\)" "gimple" } } */
+/* { dg-final { scan-tree-dump "map\\(tofrom:pt8" "gimple" } } */
+}
+
+void
+f9 (int sc9, struct S ag9)
+{
+ char ar9[sc9 + 2];
+ foo (ar9);
+ #pragma omp target defaultmap(none) map(to: ar9, ag9) firstprivate (sc9)
+ bar (sc9, ar9, ag9, &sc9);
+}
--- gcc/testsuite/c-c++-common/gomp/defaultmap-3.c.jj 2018-05-03
17:29:17.108419915 +0200
+++ gcc/testsuite/c-c++-common/gomp/defaultmap-3.c 2018-05-03
18:17:47.035727866 +0200
@@ -0,0 +1,34 @@
+/* { dg-do compile } */
+
+struct S { int s; };
+void foo (char *);
+void bar (int, char *, struct S, int *);
+#pragma omp declare target to (bar)
+#define N 16
+
+void
+f1 (int sc1, struct S ag1, int *pt1)
+{
+ char ar1[N];
+ foo (ar1);
+ #pragma omp target defaultmap(default:scalar) defaultmap(to:aggregate)
defaultmap(none:pointer) /* { dg-error "enclosing 'target'" } */
+ bar (sc1, ar1, ag1, pt1); /* { dg-error "'pt1' not specified in enclosing
'target'" } */
+}
+
+void
+f2 (int sc2, struct S ag2, int *pt2)
+{
+ char ar2[N];
+ foo (ar2);
+ #pragma omp target defaultmap(none:scalar) defaultmap(from:aggregate)
defaultmap(default:pointer) /* { dg-error "enclosing 'target'" } */
+ bar (sc2, ar2, ag2, pt2); /* { dg-error "'sc2' not specified in enclosing
'target'" } */
+}
+
+void
+f3 (int sc3, struct S ag3, int *pt3)
+{
+ char ar3[N];
+ foo (ar3);
+ #pragma omp target defaultmap(firstprivate:scalar)
defaultmap(none:aggregate) defaultmap(to:pointer) /* { dg-error "enclosing
'target'" } */
+ bar (sc3, ar3, ag3, pt3); /* { dg-error "'ar3' not specified in enclosing
'target'" } */
+} /* { dg-error "'ag3' not specified in enclosing
'target'" "" { target *-*-* } .-1 } */
Jakub
