+2015-05-19 Jakub Jelinek <jakub@redhat.com>
+
+ PR middle-end/66199
+ * tree.h (OMP_TEAMS_COMBINED): Define.
+ * gimplify.c (enum gimplify_omp_var_data): Add
+ GOVD_LINEAR_LASTPRIVATE_NO_OUTER.
+ (enum omp_region_type): Add ORT_COMBINED_TEAMS.
+ (omp_notice_variable): Accept both ORT_TEAMS
+ and ORT_COMBINED_TEAMS. Don't recurse if
+ GOVD_LINEAR_LASTPRIVATE_NO_OUTER is set and either
+ GOVD_LINEAR is set, or GOVD_LASTPRIVATE without
+ GOVD_FIRSTPRIVATE.
+ (omp_no_lastprivate): New function.
+ (gimplify_scan_omp_clauses): For OMP_CLAUSE_LASTPRIVATE
+ and OMP_CLAUSE_LINEAR, if omp_no_lastprivate, don't
+ notice_outer and set appropriate bits, otherwise make
+ sure default(none) combined constructs won't complain.
+ (gimplify_adjust_omp_clauses): Remove OMP_CLAUSE_LINEAR
+ outer special casing, for OMP_CLAUSE_LASTPRIVATE if
+ omp_no_lastprivate either remove the clause or turn it
+ into OMP_CLAUSE_PRIVATE.
+ (gimplify_omp_for): Fix up handling of implicit
+ lastprivate or linear iterators.
+ (gimplify_omp_workshare): For OMP_TEAMS_COMBINED use
+ ORT_COMBINED_TEAMS.
+ * omp-low.c (lower_omp_for_lastprivate): For combined
+ for simd use fd.loop.n2 from the for rather than simd.
+
2015-05-19 Richard Sandiford <richard.sandiford@arm.com>
* config/cris/cris.c (cris_expand_prologue): Use gen_raw_REG
+2015-05-19 Jakub Jelinek <jakub@redhat.com>
+
+ PR middle-end/66199
+ * c-parser.c (c_parser_omp_for_loop): Don't add
+ OMP_CLAUSE_SHARED to OMP_PARALLEL_CLAUSES when moving
+ OMP_CLAUSE_LASTPRIVATE clause to OMP_FOR_CLAUSES.
+ (c_parser_omp_teams): Set OMP_TEAMS_COMBINED for combined
+ constructs.
+
2015-05-19 Mikhail Maltsev <maltsevm@gmail.com>
* c-typeck.c (build_array_ref): Use std::swap instead of explicit
}
else
{
- /* Copy lastprivate (decl) clause to OMP_FOR_CLAUSES,
- change it to shared (decl) in
- OMP_PARALLEL_CLAUSES. */
- tree l = build_omp_clause (OMP_CLAUSE_LOCATION (*c),
- OMP_CLAUSE_LASTPRIVATE);
- OMP_CLAUSE_DECL (l) = OMP_CLAUSE_DECL (*c);
+ /* Move lastprivate (decl) clause to OMP_FOR_CLAUSES. */
+ tree l = *c;
+ *c = OMP_CLAUSE_CHAIN (*c);
if (code == OMP_SIMD)
{
OMP_CLAUSE_CHAIN (l)
OMP_CLAUSE_CHAIN (l) = clauses;
clauses = l;
}
- OMP_CLAUSE_SET_CODE (*c, OMP_CLAUSE_SHARED);
}
}
}
TREE_TYPE (ret) = void_type_node;
OMP_TEAMS_CLAUSES (ret) = clauses;
OMP_TEAMS_BODY (ret) = block;
+ OMP_TEAMS_COMBINED (ret) = 1;
return add_stmt (ret);
}
}
+2015-05-19 Jakub Jelinek <jakub@redhat.com>
+
+ PR middle-end/66199
+ * parser.c (cp_parser_omp_for_loop): Don't add
+ OMP_CLAUSE_SHARED to OMP_PARALLEL_CLAUSES when moving
+ OMP_CLAUSE_LASTPRIVATE clause to OMP_FOR_CLAUSES.
+ (cp_parser_omp_teams): Set OMP_TEAMS_COMBINED for combined
+ constructs.
+
2015-05-19 Mikhail Maltsev <maltsevm@gmail.com>
* typeck.c (composite_pointer_type): Use std::swap instead of explicit
else if (OMP_CLAUSE_CODE (*c) == OMP_CLAUSE_LASTPRIVATE
&& OMP_CLAUSE_DECL (*c) == real_decl)
{
- /* Add lastprivate (decl) clause to OMP_FOR_CLAUSES,
- change it to shared (decl) in OMP_PARALLEL_CLAUSES. */
- tree l = build_omp_clause (loc, OMP_CLAUSE_LASTPRIVATE);
- OMP_CLAUSE_DECL (l) = real_decl;
- CP_OMP_CLAUSE_INFO (l) = CP_OMP_CLAUSE_INFO (*c);
+ /* Move lastprivate (decl) clause to OMP_FOR_CLAUSES. */
+ tree l = *c;
+ *c = OMP_CLAUSE_CHAIN (*c);
if (code == OMP_SIMD)
{
OMP_CLAUSE_CHAIN (l) = cclauses[C_OMP_CLAUSE_SPLIT_FOR];
OMP_CLAUSE_CHAIN (l) = clauses;
clauses = l;
}
- OMP_CLAUSE_SET_CODE (*c, OMP_CLAUSE_SHARED);
- CP_OMP_CLAUSE_INFO (*c) = NULL;
add_private_clause = false;
}
else
TREE_TYPE (ret) = void_type_node;
OMP_TEAMS_CLAUSES (ret) = clauses;
OMP_TEAMS_BODY (ret) = body;
+ OMP_TEAMS_COMBINED (ret) = 1;
return add_stmt (ret);
}
}
+2015-05-19 Jakub Jelinek <jakub@redhat.com>
+
+ PR middle-end/66199
+ * trans-openmp.c (gfc_trans_omp_teams): Set OMP_TEAMS_COMBINED for
+ combined constructs.
+ (gfc_trans_omp_target): Make sure BIND_EXPR has non-NULL
+ BIND_EXPR_BLOCK.
+
2015-05-19 David Malcolm <dmalcolm@redhat.com>
* cpp.c (maybe_print_line): Strengthen local "map" from
stmtblock_t block;
gfc_omp_clauses clausesa_buf[GFC_OMP_SPLIT_NUM];
tree stmt, omp_clauses = NULL_TREE;
+ bool combined = true;
gfc_start_block (&block);
if (clausesa == NULL)
case EXEC_OMP_TARGET_TEAMS:
case EXEC_OMP_TEAMS:
stmt = gfc_trans_omp_code (code->block->next, true);
+ combined = false;
break;
case EXEC_OMP_TARGET_TEAMS_DISTRIBUTE:
case EXEC_OMP_TEAMS_DISTRIBUTE:
}
stmt = build2_loc (input_location, OMP_TEAMS, void_type_node, stmt,
omp_clauses);
+ if (combined)
+ OMP_TEAMS_COMBINED (stmt) = 1;
gfc_add_expr_to_block (&block, stmt);
return gfc_finish_block (&block);
}
if (code->op == EXEC_OMP_TARGET)
stmt = gfc_trans_omp_code (code->block->next, true);
else
- stmt = gfc_trans_omp_teams (code, clausesa);
- if (TREE_CODE (stmt) != BIND_EXPR)
- stmt = build3_v (BIND_EXPR, NULL, stmt, NULL_TREE);
+ {
+ pushlevel ();
+ stmt = gfc_trans_omp_teams (code, clausesa);
+ if (TREE_CODE (stmt) != BIND_EXPR)
+ stmt = build3_v (BIND_EXPR, NULL, stmt, poplevel (1, 0));
+ else
+ poplevel (0, 0);
+ }
if (flag_openmp)
stmt = build2_loc (input_location, OMP_TARGET, void_type_node, stmt,
omp_clauses);
/* Flag for GOVD_MAP: don't copy back. */
GOVD_MAP_TO_ONLY = 8192,
+ /* Flag for GOVD_LINEAR or GOVD_LASTPRIVATE: no outer reference. */
+ GOVD_LINEAR_LASTPRIVATE_NO_OUTER = 16384,
+
GOVD_DATA_SHARE_CLASS = (GOVD_SHARED | GOVD_PRIVATE | GOVD_FIRSTPRIVATE
| GOVD_LASTPRIVATE | GOVD_REDUCTION | GOVD_LINEAR
| GOVD_LOCAL)
ORT_TASK = 4,
ORT_UNTIED_TASK = 5,
ORT_TEAMS = 8,
+ ORT_COMBINED_TEAMS = 9,
/* Data region. */
ORT_TARGET_DATA = 16,
/* Data region with offloading. */
DECL_NAME (lang_hooks.decls.omp_report_decl (decl)));
error_at (ctx->location, "enclosing task");
}
- else if (ctx->region_type == ORT_TEAMS)
+ else if (ctx->region_type & ORT_TEAMS)
{
error ("%qE not specified in enclosing teams construct",
DECL_NAME (lang_hooks.decls.omp_report_decl (decl)));
need to propagate anything to an outer context. */
if ((flags & GOVD_PRIVATE) && !(flags & GOVD_PRIVATE_OUTER_REF))
return ret;
+ if ((flags & (GOVD_LINEAR | GOVD_LINEAR_LASTPRIVATE_NO_OUTER))
+ == (GOVD_LINEAR | GOVD_LINEAR_LASTPRIVATE_NO_OUTER))
+ return ret;
+ if ((flags & (GOVD_FIRSTPRIVATE | GOVD_LASTPRIVATE
+ | GOVD_LINEAR_LASTPRIVATE_NO_OUTER))
+ == (GOVD_LASTPRIVATE | GOVD_LINEAR_LASTPRIVATE_NO_OUTER))
+ return ret;
if (ctx->outer_context
&& omp_notice_variable (ctx->outer_context, decl, in_code))
return true;
return false;
}
+/* Return true if the CTX is combined with distribute and thus
+ lastprivate can't be supported. */
+
+static bool
+omp_no_lastprivate (struct gimplify_omp_ctx *ctx)
+{
+ do
+ {
+ if (ctx->outer_context == NULL)
+ return false;
+ ctx = ctx->outer_context;
+ switch (ctx->region_type)
+ {
+ case ORT_WORKSHARE:
+ if (!ctx->combined_loop)
+ return false;
+ if (ctx->distribute)
+ return true;
+ break;
+ case ORT_COMBINED_PARALLEL:
+ break;
+ case ORT_COMBINED_TEAMS:
+ return true;
+ default:
+ return false;
+ }
+ }
+ while (1);
+}
+
/* Scan the OMP clauses in *LIST_P, installing mappings into a new
and previous omp contexts. */
case OMP_CLAUSE_LASTPRIVATE:
flags = GOVD_LASTPRIVATE | GOVD_SEEN | GOVD_EXPLICIT;
check_non_private = "lastprivate";
+ decl = OMP_CLAUSE_DECL (c);
+ if (omp_no_lastprivate (ctx))
+ {
+ notice_outer = false;
+ flags |= GOVD_LINEAR_LASTPRIVATE_NO_OUTER;
+ }
+ else if (error_operand_p (decl))
+ goto do_add;
+ else if (outer_ctx
+ && outer_ctx->region_type == ORT_COMBINED_PARALLEL
+ && splay_tree_lookup (outer_ctx->variables,
+ (splay_tree_key) decl) == NULL)
+ omp_add_variable (outer_ctx, decl, GOVD_SHARED | GOVD_SEEN);
+ else if (outer_ctx
+ && outer_ctx->region_type == ORT_WORKSHARE
+ && outer_ctx->combined_loop
+ && splay_tree_lookup (outer_ctx->variables,
+ (splay_tree_key) decl) == NULL
+ && !omp_check_private (outer_ctx, decl, false))
+ {
+ omp_add_variable (outer_ctx, decl, GOVD_LASTPRIVATE | GOVD_SEEN);
+ if (outer_ctx->outer_context
+ && (outer_ctx->outer_context->region_type
+ == ORT_COMBINED_PARALLEL)
+ && splay_tree_lookup (outer_ctx->outer_context->variables,
+ (splay_tree_key) decl) == NULL)
+ omp_add_variable (outer_ctx->outer_context, decl,
+ GOVD_SHARED | GOVD_SEEN);
+ }
goto do_add;
case OMP_CLAUSE_REDUCTION:
flags = GOVD_REDUCTION | GOVD_SEEN | GOVD_EXPLICIT;
remove = true;
break;
}
+ else
+ {
+ /* For combined #pragma omp parallel for simd, need to put
+ lastprivate and perhaps firstprivate too on the
+ parallel. Similarly for #pragma omp for simd. */
+ struct gimplify_omp_ctx *octx = outer_ctx;
+ decl = NULL_TREE;
+ if (omp_no_lastprivate (ctx))
+ OMP_CLAUSE_LINEAR_NO_COPYOUT (c) = 1;
+ do
+ {
+ if (OMP_CLAUSE_LINEAR_NO_COPYIN (c)
+ && OMP_CLAUSE_LINEAR_NO_COPYOUT (c))
+ break;
+ decl = OMP_CLAUSE_DECL (c);
+ if (error_operand_p (decl))
+ {
+ decl = NULL_TREE;
+ break;
+ }
+ if (octx
+ && octx->region_type == ORT_WORKSHARE
+ && octx->combined_loop)
+ {
+ if (octx->outer_context
+ && (octx->outer_context->region_type
+ == ORT_COMBINED_PARALLEL
+ || (octx->outer_context->region_type
+ == ORT_COMBINED_TEAMS)))
+ octx = octx->outer_context;
+ else if (omp_check_private (octx, decl, false))
+ break;
+ }
+ else
+ break;
+ gcc_checking_assert (splay_tree_lookup (octx->variables,
+ (splay_tree_key)
+ decl) == NULL);
+ flags = GOVD_SEEN;
+ if (!OMP_CLAUSE_LINEAR_NO_COPYIN (c))
+ flags |= GOVD_FIRSTPRIVATE;
+ if (!OMP_CLAUSE_LINEAR_NO_COPYOUT (c))
+ flags |= GOVD_LASTPRIVATE;
+ omp_add_variable (octx, decl, flags);
+ if (octx->outer_context == NULL)
+ break;
+ octx = octx->outer_context;
+ }
+ while (1);
+ if (octx
+ && decl
+ && (!OMP_CLAUSE_LINEAR_NO_COPYIN (c)
+ || !OMP_CLAUSE_LINEAR_NO_COPYOUT (c)))
+ omp_notice_variable (octx, decl, true);
+ }
flags = GOVD_LINEAR | GOVD_EXPLICIT;
+ if (OMP_CLAUSE_LINEAR_NO_COPYIN (c)
+ && OMP_CLAUSE_LINEAR_NO_COPYOUT (c))
+ {
+ notice_outer = false;
+ flags |= GOVD_LINEAR_LASTPRIVATE_NO_OUTER;
+ }
goto do_add;
case OMP_CLAUSE_MAP:
OMP_CLAUSE_SET_CODE (c, OMP_CLAUSE_PRIVATE);
OMP_CLAUSE_PRIVATE_DEBUG (c) = 1;
}
- if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_LINEAR
- && ctx->outer_context
- && !(OMP_CLAUSE_LINEAR_NO_COPYIN (c)
- && OMP_CLAUSE_LINEAR_NO_COPYOUT (c)))
- {
- if (ctx->outer_context->combined_loop
- && !OMP_CLAUSE_LINEAR_NO_COPYIN (c))
- {
- n = splay_tree_lookup (ctx->outer_context->variables,
- (splay_tree_key) decl);
- if (n == NULL
- || (n->value & GOVD_DATA_SHARE_CLASS) == 0)
- {
- int flags = GOVD_FIRSTPRIVATE;
- /* #pragma omp distribute does not allow
- lastprivate clause. */
- if (!ctx->outer_context->distribute)
- flags |= GOVD_LASTPRIVATE;
- if (n == NULL)
- omp_add_variable (ctx->outer_context, decl,
- flags | GOVD_SEEN);
- else
- n->value |= flags | GOVD_SEEN;
- }
- }
- else if (!is_global_var (decl))
- omp_notice_variable (ctx->outer_context, decl, true);
- }
}
break;
n = splay_tree_lookup (ctx->variables, (splay_tree_key) decl);
OMP_CLAUSE_LASTPRIVATE_FIRSTPRIVATE (c)
= (n->value & GOVD_FIRSTPRIVATE) != 0;
+ if (omp_no_lastprivate (ctx))
+ {
+ if (OMP_CLAUSE_LASTPRIVATE_FIRSTPRIVATE (c))
+ remove = true;
+ else
+ OMP_CLAUSE_CODE (c) = OMP_CLAUSE_PRIVATE;
+ }
break;
case OMP_CLAUSE_ALIGNED:
gcc_unreachable ();
}
+ /* Set OMP_CLAUSE_LINEAR_NO_COPYIN flag on explicit linear
+ clause for the IV. */
+ if (simd && TREE_VEC_LENGTH (OMP_FOR_INIT (for_stmt)) == 1)
+ {
+ t = TREE_VEC_ELT (OMP_FOR_INIT (for_stmt), 0);
+ gcc_assert (TREE_CODE (t) == MODIFY_EXPR);
+ decl = TREE_OPERAND (t, 0);
+ for (tree c = OMP_FOR_CLAUSES (for_stmt); c; c = OMP_CLAUSE_CHAIN (c))
+ if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_LINEAR
+ && OMP_CLAUSE_DECL (c) == decl)
+ {
+ OMP_CLAUSE_LINEAR_NO_COPYIN (c) = 1;
+ break;
+ }
+ }
+
gimplify_scan_omp_clauses (&OMP_FOR_CLAUSES (for_stmt), pre_p,
simd ? ORT_SIMD : ORT_WORKSHARE);
if (TREE_CODE (for_stmt) == OMP_DISTRIBUTE)
{
c = build_omp_clause (input_location, OMP_CLAUSE_LINEAR);
OMP_CLAUSE_LINEAR_NO_COPYIN (c) = 1;
- if (has_decl_expr
- && bitmap_bit_p (has_decl_expr, DECL_UID (decl)))
- OMP_CLAUSE_LINEAR_NO_COPYOUT (c) = 1;
+ unsigned int flags = GOVD_LINEAR | GOVD_EXPLICIT | GOVD_SEEN;
+ if ((has_decl_expr
+ && bitmap_bit_p (has_decl_expr, DECL_UID (decl)))
+ || omp_no_lastprivate (gimplify_omp_ctxp))
+ {
+ OMP_CLAUSE_LINEAR_NO_COPYOUT (c) = 1;
+ flags |= GOVD_LINEAR_LASTPRIVATE_NO_OUTER;
+ }
OMP_CLAUSE_DECL (c) = decl;
OMP_CLAUSE_CHAIN (c) = OMP_FOR_CLAUSES (for_stmt);
OMP_FOR_CLAUSES (for_stmt) = c;
- omp_add_variable (gimplify_omp_ctxp, decl,
- GOVD_LINEAR | GOVD_EXPLICIT | GOVD_SEEN);
+
+ omp_add_variable (gimplify_omp_ctxp, decl, flags);
+ struct gimplify_omp_ctx *outer
+ = gimplify_omp_ctxp->outer_context;
+ if (outer && !OMP_CLAUSE_LINEAR_NO_COPYOUT (c))
+ {
+ if (outer->region_type == ORT_WORKSHARE
+ && outer->combined_loop)
+ {
+ if (outer->outer_context
+ && (outer->outer_context->region_type
+ == ORT_COMBINED_PARALLEL))
+ outer = outer->outer_context;
+ else if (omp_check_private (outer, decl, false))
+ outer = NULL;
+ }
+ else if (outer->region_type != ORT_COMBINED_PARALLEL)
+ outer = NULL;
+ if (outer)
+ {
+ omp_add_variable (outer, decl,
+ GOVD_LASTPRIVATE | GOVD_SEEN);
+ if (outer->outer_context)
+ omp_notice_variable (outer->outer_context, decl, true);
+ }
+ }
}
else
{
bool lastprivate
= (!has_decl_expr
- || !bitmap_bit_p (has_decl_expr, DECL_UID (decl)));
- if (lastprivate
- && gimplify_omp_ctxp->outer_context
- && gimplify_omp_ctxp->outer_context->region_type
- == ORT_WORKSHARE
- && gimplify_omp_ctxp->outer_context->combined_loop
- && !gimplify_omp_ctxp->outer_context->distribute)
+ || !bitmap_bit_p (has_decl_expr, DECL_UID (decl)))
+ && !omp_no_lastprivate (gimplify_omp_ctxp);
+ struct gimplify_omp_ctx *outer
+ = gimplify_omp_ctxp->outer_context;
+ if (outer && lastprivate)
{
- struct gimplify_omp_ctx *outer
- = gimplify_omp_ctxp->outer_context;
- n = splay_tree_lookup (outer->variables,
- (splay_tree_key) decl);
- if (n != NULL
- && (n->value & GOVD_DATA_SHARE_CLASS) == GOVD_LOCAL)
- lastprivate = false;
- else if (omp_check_private (outer, decl, false))
- error ("lastprivate variable %qE is private in outer "
- "context", DECL_NAME (decl));
- else
+ if (outer->region_type == ORT_WORKSHARE
+ && outer->combined_loop)
+ {
+ if (outer->outer_context
+ && (outer->outer_context->region_type
+ == ORT_COMBINED_PARALLEL))
+ outer = outer->outer_context;
+ else if (omp_check_private (outer, decl, false))
+ outer = NULL;
+ }
+ else if (outer->region_type != ORT_COMBINED_PARALLEL)
+ outer = NULL;
+ if (outer)
{
omp_add_variable (outer, decl,
GOVD_LASTPRIVATE | GOVD_SEEN);
omp_notice_variable (outer->outer_context, decl, true);
}
}
+
c = build_omp_clause (input_location,
lastprivate ? OMP_CLAUSE_LASTPRIVATE
: OMP_CLAUSE_PRIVATE);
ort = ORT_TARGET_DATA;
break;
case OMP_TEAMS:
- ort = ORT_TEAMS;
+ ort = OMP_TEAMS_COMBINED (expr) ? ORT_COMBINED_TEAMS : ORT_TEAMS;
break;
default:
gcc_unreachable ();
cond_code = EQ_EXPR;
}
- cond = build2 (cond_code, boolean_type_node, fd->loop.v, fd->loop.n2);
+ tree n2 = fd->loop.n2;
+ if (fd->collapse > 1
+ && TREE_CODE (n2) != INTEGER_CST
+ && gimple_omp_for_combined_into_p (fd->for_stmt)
+ && gimple_code (ctx->outer->stmt) == GIMPLE_OMP_FOR)
+ {
+ gomp_for *gfor = as_a <gomp_for *> (ctx->outer->stmt);
+ if (gimple_omp_for_kind (gfor) == GF_OMP_FOR_KIND_FOR)
+ {
+ struct omp_for_data outer_fd;
+ extract_omp_for_data (gfor, &outer_fd, NULL);
+ n2 = fold_convert (TREE_TYPE (n2), outer_fd.loop.n2);
+ }
+ }
+ cond = build2 (cond_code, boolean_type_node, fd->loop.v, n2);
clauses = gimple_omp_for_clauses (fd->for_stmt);
stmts = NULL;
#define OMP_PARALLEL_COMBINED(NODE) \
(OMP_PARALLEL_CHECK (NODE)->base.private_flag)
+/* True on an OMP_TEAMS statement if it represents an explicit
+ combined teams distribute constructs. */
+#define OMP_TEAMS_COMBINED(NODE) \
+ (OMP_TEAMS_CHECK (NODE)->base.private_flag)
+
/* True if OMP_ATOMIC* is supposed to be sequentially consistent
as opposed to relaxed. */
#define OMP_ATOMIC_SEQ_CST(NODE) \
+2015-05-19 Jakub Jelinek <jakub@redhat.com>
+
+ PR middle-end/66199
+ * testsuite/libgomp.c/pr66199-1.c: New test.
+ * testsuite/libgomp.c/pr66199-2.c: New test.
+ * testsuite/libgomp.c++/pr66199-1.C: New test.
+ * testsuite/libgomp.c++/pr66199-2.C: New test.
+ * testsuite/libgomp.fortran/pr66199-1.f90: New test.
+ * testsuite/libgomp.fortran/pr66199-2.f90: New test.
+
2015-05-19 Julian Brown <julian@codesourcery.com>
* plugin/plugin-nvptx.c (nvptx_get_num_devices): Return zero
--- /dev/null
+// PR middle-end/66199
+// { dg-do run }
+// { dg-options "-O2 -fopenmp" }
+
+#include "../libgomp.c/pr66199-1.c"
--- /dev/null
+// PR middle-end/66199
+// { dg-do run }
+// { dg-options "-O2 -fopenmp" }
+
+#include "../libgomp.c/pr66199-2.c"
--- /dev/null
+/* PR middle-end/66199 */
+/* { dg-do run } */
+/* { dg-options "-O2 -fopenmp" } */
+
+int u[1024], v[1024], w[1024];
+
+__attribute__((noinline, noclone)) long
+f1 (long a, long b)
+{
+ long d;
+ #pragma omp parallel for simd default(none) firstprivate (a, b) shared(u, v, w)
+ for (d = a; d < b; d++)
+ u[d] = v[d] + w[d];
+ return d;
+}
+
+__attribute__((noinline, noclone)) long
+f2 (long a, long b, long c)
+{
+ long d, e;
+ #pragma omp parallel for simd default(none) firstprivate (a, b) shared(u, v, w) linear(d) linear(c:5) lastprivate(e)
+ for (d = a; d < b; d++)
+ {
+ u[d] = v[d] + w[d];
+ c += 5;
+ e = c;
+ }
+ return d + c + e;
+}
+
+__attribute__((noinline, noclone)) long
+f3 (long a1, long b1, long a2, long b2)
+{
+ long d1, d2;
+ #pragma omp parallel for simd default(none) firstprivate (a1, b1, a2, b2) shared(u, v, w) lastprivate(d1, d2) collapse(2)
+ for (d1 = a1; d1 < b1; d1++)
+ for (d2 = a2; d2 < b2; d2++)
+ u[d1 * 32 + d2] = v[d1 * 32 + d2] + w[d1 * 32 + d2];
+ return d1 + d2;
+}
+
+__attribute__((noinline, noclone)) long
+f4 (long a1, long b1, long a2, long b2)
+{
+ long d1, d2;
+ #pragma omp parallel for simd default(none) firstprivate (a1, b1, a2, b2) shared(u, v, w) collapse(2)
+ for (d1 = a1; d1 < b1; d1++)
+ for (d2 = a2; d2 < b2; d2++)
+ u[d1 * 32 + d2] = v[d1 * 32 + d2] + w[d1 * 32 + d2];
+ return d1 + d2;
+}
+
+int
+main ()
+{
+ if (f1 (0, 1024) != 1024
+ || f2 (0, 1024, 17) != 1024 + 2 * (17 + 5 * 1024)
+ || f3 (0, 32, 0, 32) != 64
+ || f4 (0, 32, 0, 32) != 64)
+ __builtin_abort ();
+ return 0;
+}
--- /dev/null
+/* PR middle-end/66199 */
+/* { dg-do run } */
+/* { dg-options "-O2 -fopenmp" } */
+
+#pragma omp declare target
+int u[1024], v[1024], w[1024];
+#pragma omp end declare target
+
+__attribute__((noinline, noclone)) void
+f1 (long a, long b)
+{
+ long d;
+ #pragma omp target teams distribute parallel for simd default(none) firstprivate (a, b) shared(u, v, w)
+ for (d = a; d < b; d++)
+ u[d] = v[d] + w[d];
+}
+
+__attribute__((noinline, noclone)) void
+f2 (long a, long b, long c)
+{
+ long d, e;
+ #pragma omp target teams distribute parallel for simd default(none) firstprivate (a, b) shared(u, v, w) linear(d) linear(c:5) lastprivate(e)
+ for (d = a; d < b; d++)
+ {
+ u[d] = v[d] + w[d];
+ c += 5;
+ e = c;
+ }
+}
+
+__attribute__((noinline, noclone)) void
+f3 (long a1, long b1, long a2, long b2)
+{
+ long d1, d2;
+ #pragma omp target teams distribute parallel for simd default(none) firstprivate (a1, b1, a2, b2) shared(u, v, w) lastprivate(d1, d2) collapse(2)
+ for (d1 = a1; d1 < b1; d1++)
+ for (d2 = a2; d2 < b2; d2++)
+ u[d1 * 32 + d2] = v[d1 * 32 + d2] + w[d1 * 32 + d2];
+}
+
+__attribute__((noinline, noclone)) void
+f4 (long a1, long b1, long a2, long b2)
+{
+ long d1, d2;
+ #pragma omp target teams distribute parallel for simd default(none) firstprivate (a1, b1, a2, b2) shared(u, v, w) collapse(2)
+ for (d1 = a1; d1 < b1; d1++)
+ for (d2 = a2; d2 < b2; d2++)
+ u[d1 * 32 + d2] = v[d1 * 32 + d2] + w[d1 * 32 + d2];
+}
+
+int
+main ()
+{
+ f1 (0, 1024);
+ f2 (0, 1024, 17);
+ f3 (0, 32, 0, 32);
+ f4 (0, 32, 0, 32);
+ return 0;
+}
--- /dev/null
+! PR middle-end/66199
+! { dg-do run }
+! { dg-options "-O2 -fopenmp" }
+
+ integer :: u(1024), v(1024), w(1024), a, b, c, d, e, a1, b1, a2, b2, d1, d2
+ a = 1
+ b = 1024
+ d = 75
+ !$omp parallel do simd default(none) firstprivate (a, b) shared(u, v, w)
+ do d = a, b
+ u(d) = v(d) + w(d)
+ end do
+ if (d .ne. 1025) call abort
+ c = 17
+ d = 75
+ !$omp parallel do simd default(none) firstprivate (a, b) shared(u, v, w) &
+ !$omp& linear(d) linear(c:5) lastprivate(e)
+ do d = a, b
+ u(d) = v(d) + w(d)
+ c = c + 5
+ e = c
+ end do
+ if (d .ne. 1025 .or. c .ne. (17 + 5 * 1024)) call abort
+ if (e .ne. (17 + 5 * 1024)) call abort
+ a1 = 0
+ a2 = 0
+ b1 = 31
+ b2 = 31
+ d1 = 7
+ d2 = 9
+ !$omp parallel do simd default(none) firstprivate (a1, b1, a2, b2) &
+ !$omp& shared(u, v, w) lastprivate(d1, d2) collapse(2)
+ do d1 = a1, b1
+ do d2 = a2, b2
+ u(d1 * 32 + d2 + 1) = v(d1 * 32 + d2 + 1) + w(d1 * 32 + d2 + 1)
+ end do
+ end do
+ if (d1 .ne. 32 .or. d2 .ne. 32) call abort
+ d1 = 7
+ d2 = 9
+ !$omp parallel do simd default(none) firstprivate (a1, b1, a2, b2) &
+ !$omp& shared(u, v, w) collapse(2)
+ do d1 = a1, b1
+ do d2 = a2, b2
+ u(d1 * 32 + d2 + 1) = v(d1 * 32 + d2 + 1) + w(d1 * 32 + d2 + 1)
+ end do
+ end do
+ if (d1 .ne. 32 .or. d2 .ne. 32) call abort
+end
--- /dev/null
+! PR middle-end/66199
+! { dg-do run }
+! { dg-options "-O2 -fopenmp" }
+
+ integer :: u(1024), v(1024), w(1024), a, b, c, d, e, a1, b1, a2, b2, d1, d2
+ a = 1
+ b = 1024
+ d = 75
+ !$omp target teams distribute parallel do simd default(none) &
+ !$omp& firstprivate (a, b) shared(u, v, w)
+ do d = a, b
+ u(d) = v(d) + w(d)
+ end do
+ c = 17
+ d = 75
+ !$omp target teams distribute parallel do simd default(none) &
+ !$omp& firstprivate (a, b) shared(u, v, w) &
+ !$omp& linear(d) linear(c:5) lastprivate(e)
+ do d = a, b
+ u(d) = v(d) + w(d)
+ c = c + 5
+ e = c
+ end do
+ a1 = 0
+ a2 = 0
+ b1 = 31
+ b2 = 31
+ d1 = 7
+ d2 = 9
+ !$omp target teams distribute parallel do simd default(none) &
+ !$omp& firstprivate (a1, b1, a2, b2) &
+ !$omp& shared(u, v, w) lastprivate(d1, d2) collapse(2)
+ do d1 = a1, b1
+ do d2 = a2, b2
+ u(d1 * 32 + d2 + 1) = v(d1 * 32 + d2 + 1) + w(d1 * 32 + d2 + 1)
+ end do
+ end do
+ d1 = 7
+ d2 = 9
+ !$omp target teams distribute parallel do simd default(none) &
+ !$omp& firstprivate (a1, b1, a2, b2) shared(u, v, w) collapse(2)
+ do d1 = a1, b1
+ do d2 = a2, b2
+ u(d1 * 32 + d2 + 1) = v(d1 * 32 + d2 + 1) + w(d1 * 32 + d2 + 1)
+ end do
+ end do
+end
projects / gcc.git / commitdiff