@@ -251,6 +251,7 @@ DEF_FUNCTION_TYPE_1 (BT_FN_INT_CONST_STR
DEF_FUNCTION_TYPE_1 (BT_FN_PTR_PTR, BT_PTR, BT_PTR)
DEF_FUNCTION_TYPE_1 (BT_FN_VOID_VALIST_REF, BT_VOID, BT_VALIST_REF)
DEF_FUNCTION_TYPE_1 (BT_FN_VOID_INT, BT_VOID, BT_INT)
+DEF_FUNCTION_TYPE_1 (BT_FN_VOID_BOOL, BT_VOID, BT_BOOL)
DEF_FUNCTION_TYPE_1 (BT_FN_FLOAT_CONST_STRING, BT_FLOAT, BT_CONST_STRING)
DEF_FUNCTION_TYPE_1 (BT_FN_DOUBLE_CONST_STRING, BT_DOUBLE, BT_CONST_STRING)
DEF_FUNCTION_TYPE_1 (BT_FN_LONGDOUBLE_CONST_STRING,
@@ -621,6 +622,9 @@ DEF_FUNCTION_TYPE_3 (BT_FN_VOID_UINT32_U
BT_VOID, BT_UINT32, BT_UINT64, BT_PTR)
DEF_FUNCTION_TYPE_3 (BT_FN_VOID_UINT32_UINT32_PTR,
BT_VOID, BT_UINT32, BT_UINT32, BT_PTR)
+DEF_FUNCTION_TYPE_3 (BT_FN_VOID_SIZE_SIZE_PTR, BT_VOID, BT_SIZE, BT_SIZE,
+ BT_PTR)
+DEF_FUNCTION_TYPE_3 (BT_FN_UINT_UINT_PTR_PTR, BT_UINT, BT_UINT, BT_PTR, BT_PTR)
DEF_FUNCTION_TYPE_4 (BT_FN_SIZE_CONST_PTR_SIZE_SIZE_FILEPTR,
BT_SIZE, BT_CONST_PTR, BT_SIZE, BT_SIZE, BT_FILEPTR)
@@ -644,6 +648,8 @@ DEF_FUNCTION_TYPE_4 (BT_FN_INT_FILEPTR_I
BT_INT, BT_FILEPTR, BT_INT, BT_CONST_STRING, BT_VALIST_ARG)
DEF_FUNCTION_TYPE_4 (BT_FN_VOID_OMPFN_PTR_UINT_UINT,
BT_VOID, BT_PTR_FN_VOID_PTR, BT_PTR, BT_UINT, BT_UINT)
+DEF_FUNCTION_TYPE_4 (BT_FN_UINT_OMPFN_PTR_UINT_UINT,
+ BT_UINT, BT_PTR_FN_VOID_PTR, BT_PTR, BT_UINT, BT_UINT)
DEF_FUNCTION_TYPE_4 (BT_FN_VOID_PTR_WORD_WORD_PTR,
BT_VOID, BT_PTR, BT_WORD, BT_WORD, BT_PTR)
DEF_FUNCTION_TYPE_4 (BT_FN_VOID_SIZE_VPTR_PTR_INT, BT_VOID, BT_SIZE,
@@ -729,6 +735,12 @@ DEF_FUNCTION_TYPE_7 (BT_FN_VOID_INT_SIZE
DEF_FUNCTION_TYPE_8 (BT_FN_VOID_OMPFN_PTR_UINT_LONG_LONG_LONG_LONG_UINT,
BT_VOID, BT_PTR_FN_VOID_PTR, BT_PTR, BT_UINT,
BT_LONG, BT_LONG, BT_LONG, BT_LONG, BT_UINT)
+DEF_FUNCTION_TYPE_8 (BT_FN_BOOL_UINT_LONGPTR_LONG_LONG_LONGPTR_LONGPTR_PTR_PTR,
+ BT_BOOL, BT_UINT, BT_PTR_LONG, BT_LONG, BT_LONG,
+ BT_PTR_LONG, BT_PTR_LONG, BT_PTR, BT_PTR)
+DEF_FUNCTION_TYPE_8 (BT_FN_BOOL_UINT_ULLPTR_LONG_ULL_ULLPTR_ULLPTR_PTR_PTR,
+ BT_BOOL, BT_UINT, BT_PTR_ULONGLONG, BT_LONG, BT_ULONGLONG,
+ BT_PTR_ULONGLONG, BT_PTR_ULONGLONG, BT_PTR, BT_PTR)
DEF_FUNCTION_TYPE_9 (BT_FN_VOID_OMPFN_PTR_OMPCPYFN_LONG_LONG_BOOL_UINT_PTR_INT,
BT_VOID, BT_PTR_FN_VOID_PTR, BT_PTR,
@@ -737,6 +749,14 @@ DEF_FUNCTION_TYPE_9 (BT_FN_VOID_OMPFN_PT
DEF_FUNCTION_TYPE_9 (BT_FN_VOID_INT_OMPFN_SIZE_PTR_PTR_PTR_UINT_PTR_PTR,
BT_VOID, BT_INT, BT_PTR_FN_VOID_PTR, BT_SIZE, BT_PTR,
BT_PTR, BT_PTR, BT_UINT, BT_PTR, BT_PTR)
+DEF_FUNCTION_TYPE_9 (BT_FN_BOOL_LONG_LONG_LONG_LONG_LONG_LONGPTR_LONGPTR_PTR_PTR,
+ BT_BOOL, BT_LONG, BT_LONG, BT_LONG, BT_LONG, BT_LONG,
+ BT_PTR_LONG, BT_PTR_LONG, BT_PTR, BT_PTR)
+
+DEF_FUNCTION_TYPE_10 (BT_FN_BOOL_BOOL_ULL_ULL_ULL_LONG_ULL_ULLPTR_ULLPTR_PTR_PTR,
+ BT_BOOL, BT_BOOL, BT_ULONGLONG, BT_ULONGLONG,
+ BT_ULONGLONG, BT_LONG, BT_ULONGLONG, BT_PTR_ULONGLONG,
+ BT_PTR_ULONGLONG, BT_PTR, BT_PTR)
DEF_FUNCTION_TYPE_11 (BT_FN_VOID_OMPFN_PTR_OMPCPYFN_LONG_LONG_UINT_LONG_INT_LONG_LONG_LONG,
BT_VOID, BT_PTR_FN_VOID_PTR, BT_PTR,
@@ -1724,7 +1724,8 @@ open_base_files (void)
"tree-dfa.h", "tree-ssa.h", "reload.h", "cpplib.h", "tree-chrec.h",
"except.h", "output.h", "cfgloop.h", "target.h", "lto-streamer.h",
"target-globals.h", "ipa-ref.h", "cgraph.h", "symbol-summary.h",
- "ipa-prop.h", "ipa-fnsummary.h", "dwarf2out.h", "omp-offload.h", NULL
+ "ipa-prop.h", "ipa-fnsummary.h", "dwarf2out.h", "omp-general.h",
+ "omp-offload.h", NULL
};
const char *const *ifp;
outf_p gtype_desc_c;
@@ -924,7 +924,7 @@ gimple_build_omp_critical (gimple_seq bo
BODY is sequence of statements inside the for loop.
KIND is the `for' variant.
- CLAUSES, are any of the construct's clauses.
+ CLAUSES are any of the construct's clauses.
COLLAPSE is the collapse count.
PRE_BODY is the sequence of statements that are loop invariant. */
@@ -950,7 +950,7 @@ gimple_build_omp_for (gimple_seq body, i
/* Build a GIMPLE_OMP_PARALLEL statement.
BODY is sequence of statements which are executed in parallel.
- CLAUSES, are the OMP parallel construct's clauses.
+ CLAUSES are the OMP parallel construct's clauses.
CHILD_FN is the function created for the parallel threads to execute.
DATA_ARG are the shared data argument(s). */
@@ -973,7 +973,7 @@ gimple_build_omp_parallel (gimple_seq bo
/* Build a GIMPLE_OMP_TASK statement.
BODY is sequence of statements which are executed by the explicit task.
- CLAUSES, are the OMP parallel construct's clauses.
+ CLAUSES are the OMP task construct's clauses.
CHILD_FN is the function created for the parallel threads to execute.
DATA_ARG are the shared data argument(s).
COPY_FN is the optional function for firstprivate initialization.
@@ -1044,12 +1044,14 @@ gimple_build_omp_grid_body (gimple_seq b
/* Build a GIMPLE_OMP_TASKGROUP statement.
BODY is the sequence of statements to be executed by the taskgroup
- construct. */
+ construct.
+ CLAUSES are any of the construct's clauses. */
gimple *
-gimple_build_omp_taskgroup (gimple_seq body)
+gimple_build_omp_taskgroup (gimple_seq body, tree clauses)
{
gimple *p = gimple_alloc (GIMPLE_OMP_TASKGROUP, 0);
+ gimple_omp_taskgroup_set_clauses (p, clauses);
if (body)
gimple_omp_set_body (p, body);
@@ -1192,12 +1194,13 @@ gimple_build_omp_teams (gimple_seq body,
/* Build a GIMPLE_OMP_ATOMIC_LOAD statement. */
gomp_atomic_load *
-gimple_build_omp_atomic_load (tree lhs, tree rhs)
+gimple_build_omp_atomic_load (tree lhs, tree rhs, enum omp_memory_order mo)
{
gomp_atomic_load *p
= as_a <gomp_atomic_load *> (gimple_alloc (GIMPLE_OMP_ATOMIC_LOAD, 0));
gimple_omp_atomic_load_set_lhs (p, lhs);
gimple_omp_atomic_load_set_rhs (p, rhs);
+ gimple_omp_atomic_set_memory_order (p, mo);
return p;
}
@@ -1206,11 +1209,12 @@ gimple_build_omp_atomic_load (tree lhs,
VAL is the value we are storing. */
gomp_atomic_store *
-gimple_build_omp_atomic_store (tree val)
+gimple_build_omp_atomic_store (tree val, enum omp_memory_order mo)
{
gomp_atomic_store *p
= as_a <gomp_atomic_store *> (gimple_alloc (GIMPLE_OMP_ATOMIC_STORE, 0));
gimple_omp_atomic_store_set_val (p, val);
+ gimple_omp_atomic_set_memory_order (p, mo);
return p;
}
@@ -1935,6 +1939,11 @@ gimple_copy (gimple *stmt)
gimple_omp_ordered_set_clauses (as_a <gomp_ordered *> (copy), t);
goto copy_omp_body;
+ case GIMPLE_OMP_TASKGROUP:
+ t = unshare_expr (gimple_omp_taskgroup_clauses (stmt));
+ gimple_omp_taskgroup_set_clauses (copy, t);
+ goto copy_omp_body;
+
case GIMPLE_OMP_SECTIONS:
t = unshare_expr (gimple_omp_sections_clauses (stmt));
gimple_omp_sections_set_clauses (copy, t);
@@ -1971,7 +1980,6 @@ gimple_copy (gimple *stmt)
case GIMPLE_OMP_SECTION:
case GIMPLE_OMP_MASTER:
- case GIMPLE_OMP_TASKGROUP:
case GIMPLE_OMP_GRID_BODY:
copy_omp_body:
new_seq = gimple_seq_copy (gimple_omp_body (stmt));
@@ -279,9 +279,10 @@ DEFGSCODE(GIMPLE_OMP_FOR, "gimple_omp_fo
BODY is the sequence of statements to execute in the master section. */
DEFGSCODE(GIMPLE_OMP_MASTER, "gimple_omp_master", GSS_OMP)
-/* GIMPLE_OMP_TASKGROUP <BODY> represents #pragma omp taskgroup.
- BODY is the sequence of statements to execute in the taskgroup section. */
-DEFGSCODE(GIMPLE_OMP_TASKGROUP, "gimple_omp_taskgroup", GSS_OMP)
+/* GIMPLE_OMP_TASKGROUP <BODY, CLAUSES> represents #pragma omp taskgroup.
+ BODY is the sequence of statements inside the taskgroup section.
+ CLAUSES is an OMP_CLAUSE chain holding the associated clauses. */
+DEFGSCODE(GIMPLE_OMP_TASKGROUP, "gimple_omp_taskgroup", GSS_OMP_SINGLE_LAYOUT)
/* GIMPLE_OMP_PARALLEL <BODY, CLAUSES, CHILD_FN, DATA_ARG> represents
@@ -366,10 +367,12 @@ DEFGSCODE(GIMPLE_OMP_SINGLE, "gimple_omp
implement the MAP clauses. */
DEFGSCODE(GIMPLE_OMP_TARGET, "gimple_omp_target", GSS_OMP_PARALLEL_LAYOUT)
-/* GIMPLE_OMP_TEAMS <BODY, CLAUSES> represents #pragma omp teams
+/* GIMPLE_OMP_TEAMS <BODY, CLAUSES, CHILD_FN, DATA_ARG> represents
+ #pragma omp teams
BODY is the sequence of statements inside the single section.
- CLAUSES is an OMP_CLAUSE chain holding the associated clauses. */
-DEFGSCODE(GIMPLE_OMP_TEAMS, "gimple_omp_teams", GSS_OMP_SINGLE_LAYOUT)
+ CLAUSES is an OMP_CLAUSE chain holding the associated clauses.
+ CHILD_FN and DATA_ARG like for GIMPLE_OMP_PARALLEL. */
+DEFGSCODE(GIMPLE_OMP_TEAMS, "gimple_omp_teams", GSS_OMP_PARALLEL_LAYOUT)
/* GIMPLE_OMP_ORDERED <BODY, CLAUSES> represents #pragma omp ordered.
BODY is the sequence of statements to execute in the ordered section.
@@ -151,6 +151,7 @@ enum gf_mask {
GF_OMP_PARALLEL_COMBINED = 1 << 0,
GF_OMP_PARALLEL_GRID_PHONY = 1 << 1,
GF_OMP_TASK_TASKLOOP = 1 << 0,
+ GF_OMP_TASK_TASKWAIT = 1 << 1,
GF_OMP_FOR_KIND_MASK = (1 << 4) - 1,
GF_OMP_FOR_KIND_FOR = 0,
GF_OMP_FOR_KIND_DISTRIBUTE = 1,
@@ -183,6 +184,7 @@ enum gf_mask {
GF_OMP_TARGET_KIND_OACC_DECLARE = 10,
GF_OMP_TARGET_KIND_OACC_HOST_DATA = 11,
GF_OMP_TEAMS_GRID_PHONY = 1 << 0,
+ GF_OMP_TEAMS_HOST = 1 << 1,
/* True on an GIMPLE_OMP_RETURN statement if the return does not require
a thread synchronization via some sort of barrier. The exact barrier
@@ -191,8 +193,8 @@ enum gf_mask {
GF_OMP_RETURN_NOWAIT = 1 << 0,
GF_OMP_SECTION_LAST = 1 << 0,
- GF_OMP_ATOMIC_NEED_VALUE = 1 << 0,
- GF_OMP_ATOMIC_SEQ_CST = 1 << 1,
+ GF_OMP_ATOMIC_MEMORY_ORDER = (1 << 3) - 1,
+ GF_OMP_ATOMIC_NEED_VALUE = 1 << 3,
GF_PREDICT_TAKEN = 1 << 15
};
@@ -637,7 +639,7 @@ struct GTY((tag("GSS_OMP_FOR")))
};
-/* GIMPLE_OMP_PARALLEL, GIMPLE_OMP_TARGET, GIMPLE_OMP_TASK */
+/* GIMPLE_OMP_PARALLEL, GIMPLE_OMP_TARGET, GIMPLE_OMP_TASK, GIMPLE_OMP_TEAMS */
struct GTY((tag("GSS_OMP_PARALLEL_LAYOUT")))
gimple_statement_omp_parallel_layout : public gimple_statement_omp
@@ -663,7 +665,8 @@ struct GTY((tag("GSS_OMP_PARALLEL_LAYOUT
{
/* No extra fields; adds invariant:
stmt->code == GIMPLE_OMP_PARALLEL
- || stmt->code == GIMPLE_OMP_TASK. */
+ || stmt->code == GIMPLE_OMP_TASK
+ || stmt->code == GIMPLE_OMP_TEAMS. */
};
/* GIMPLE_OMP_PARALLEL */
@@ -737,7 +740,7 @@ struct GTY((tag("GSS_OMP_CONTINUE")))
tree control_use;
};
-/* GIMPLE_OMP_SINGLE, GIMPLE_OMP_TEAMS, GIMPLE_OMP_ORDERED */
+/* GIMPLE_OMP_SINGLE, GIMPLE_OMP_ORDERED, GIMPLE_OMP_TASKGROUP. */
struct GTY((tag("GSS_OMP_SINGLE_LAYOUT")))
gimple_statement_omp_single_layout : public gimple_statement_omp
@@ -755,8 +758,8 @@ struct GTY((tag("GSS_OMP_SINGLE_LAYOUT")
stmt->code == GIMPLE_OMP_SINGLE. */
};
-struct GTY((tag("GSS_OMP_SINGLE_LAYOUT")))
- gomp_teams : public gimple_statement_omp_single_layout
+struct GTY((tag("GSS_OMP_PARALLEL_LAYOUT")))
+ gomp_teams : public gimple_statement_omp_taskreg
{
/* No extra fields; adds invariant:
stmt->code == GIMPLE_OMP_TEAMS. */
@@ -1121,7 +1124,9 @@ template <>
inline bool
is_a_helper <gimple_statement_omp_taskreg *>::test (gimple *gs)
{
- return gs->code == GIMPLE_OMP_PARALLEL || gs->code == GIMPLE_OMP_TASK;
+ return (gs->code == GIMPLE_OMP_PARALLEL
+ || gs->code == GIMPLE_OMP_TASK
+ || gs->code == GIMPLE_OMP_TEAMS);
}
template <>
@@ -1337,7 +1342,9 @@ template <>
inline bool
is_a_helper <const gimple_statement_omp_taskreg *>::test (const gimple *gs)
{
- return gs->code == GIMPLE_OMP_PARALLEL || gs->code == GIMPLE_OMP_TASK;
+ return (gs->code == GIMPLE_OMP_PARALLEL
+ || gs->code == GIMPLE_OMP_TASK
+ || gs->code == GIMPLE_OMP_TEAMS);
}
template <>
@@ -1463,7 +1470,7 @@ gomp_task *gimple_build_omp_task (gimple
gimple *gimple_build_omp_section (gimple_seq);
gimple *gimple_build_omp_master (gimple_seq);
gimple *gimple_build_omp_grid_body (gimple_seq);
-gimple *gimple_build_omp_taskgroup (gimple_seq);
+gimple *gimple_build_omp_taskgroup (gimple_seq, tree);
gomp_continue *gimple_build_omp_continue (tree, tree);
gomp_ordered *gimple_build_omp_ordered (gimple_seq, tree);
gimple *gimple_build_omp_return (bool);
@@ -1472,8 +1479,9 @@ gimple *gimple_build_omp_sections_switch
gomp_single *gimple_build_omp_single (gimple_seq, tree);
gomp_target *gimple_build_omp_target (gimple_seq, int, tree);
gomp_teams *gimple_build_omp_teams (gimple_seq, tree);
-gomp_atomic_load *gimple_build_omp_atomic_load (tree, tree);
-gomp_atomic_store *gimple_build_omp_atomic_store (tree);
+gomp_atomic_load *gimple_build_omp_atomic_load (tree, tree,
+ enum omp_memory_order);
+gomp_atomic_store *gimple_build_omp_atomic_store (tree, enum omp_memory_order);
gtransaction *gimple_build_transaction (gimple_seq);
extern void gimple_seq_add_stmt (gimple_seq *, gimple *);
extern void gimple_seq_add_stmt_without_update (gimple_seq *, gimple *);
@@ -2193,7 +2201,7 @@ static inline unsigned
gimple_omp_subcode (const gimple *s)
{
gcc_gimple_checking_assert (gimple_code (s) >= GIMPLE_OMP_ATOMIC_LOAD
- && gimple_code (s) <= GIMPLE_OMP_TEAMS);
+ && gimple_code (s) <= GIMPLE_OMP_TEAMS);
return s->subcode;
}
@@ -2331,26 +2339,27 @@ gimple_omp_atomic_set_need_value (gimple
}
-/* Return true if OMP atomic load/store statement G has the
- GF_OMP_ATOMIC_SEQ_CST flag set. */
+/* Return the memory order of the OMP atomic load/store statement G. */
-static inline bool
-gimple_omp_atomic_seq_cst_p (const gimple *g)
+static inline enum omp_memory_order
+gimple_omp_atomic_memory_order (const gimple *g)
{
if (gimple_code (g) != GIMPLE_OMP_ATOMIC_LOAD)
GIMPLE_CHECK (g, GIMPLE_OMP_ATOMIC_STORE);
- return (gimple_omp_subcode (g) & GF_OMP_ATOMIC_SEQ_CST) != 0;
+ return (enum omp_memory_order)
+ (gimple_omp_subcode (g) & GF_OMP_ATOMIC_MEMORY_ORDER);
}
-/* Set the GF_OMP_ATOMIC_SEQ_CST flag on G. */
+/* Set the memory order on G. */
static inline void
-gimple_omp_atomic_set_seq_cst (gimple *g)
+gimple_omp_atomic_set_memory_order (gimple *g, enum omp_memory_order mo)
{
if (gimple_code (g) != GIMPLE_OMP_ATOMIC_LOAD)
GIMPLE_CHECK (g, GIMPLE_OMP_ATOMIC_STORE);
- g->subcode |= GF_OMP_ATOMIC_SEQ_CST;
+ g->subcode = ((g->subcode & ~GF_OMP_ATOMIC_MEMORY_ORDER)
+ | (mo & GF_OMP_ATOMIC_MEMORY_ORDER));
}
@@ -4915,6 +4924,40 @@ gimple_omp_ordered_set_clauses (gomp_ord
}
+/* Return the clauses associated with OMP_TASKGROUP statement GS. */
+
+static inline tree
+gimple_omp_taskgroup_clauses (const gimple *gs)
+{
+ GIMPLE_CHECK (gs, GIMPLE_OMP_TASKGROUP);
+ return
+ static_cast <const gimple_statement_omp_single_layout *> (gs)->clauses;
+}
+
+
+/* Return a pointer to the clauses associated with OMP taskgroup statement
+ GS. */
+
+static inline tree *
+gimple_omp_taskgroup_clauses_ptr (gimple *gs)
+{
+ GIMPLE_CHECK (gs, GIMPLE_OMP_TASKGROUP);
+ return &static_cast <gimple_statement_omp_single_layout *> (gs)->clauses;
+}
+
+
+/* Set CLAUSES to be the clauses associated with OMP taskgroup statement
+ GS. */
+
+static inline void
+gimple_omp_taskgroup_set_clauses (gimple *gs, tree clauses)
+{
+ GIMPLE_CHECK (gs, GIMPLE_OMP_TASKGROUP);
+ static_cast <gimple_statement_omp_single_layout *> (gs)->clauses
+ = clauses;
+}
+
+
/* Return the kind of the OMP_FOR statemement G. */
static inline int
@@ -5441,6 +5484,31 @@ gimple_omp_task_set_taskloop_p (gimple *
}
+/* Return true if OMP task statement G has the
+ GF_OMP_TASK_TASKWAIT flag set. */
+
+static inline bool
+gimple_omp_task_taskwait_p (const gimple *g)
+{
+ GIMPLE_CHECK (g, GIMPLE_OMP_TASK);
+ return (gimple_omp_subcode (g) & GF_OMP_TASK_TASKWAIT) != 0;
+}
+
+
+/* Set the GF_OMP_TASK_TASKWAIT field in G depending on the boolean
+ value of TASKWAIT_P. */
+
+static inline void
+gimple_omp_task_set_taskwait_p (gimple *g, bool taskwait_p)
+{
+ GIMPLE_CHECK (g, GIMPLE_OMP_TASK);
+ if (taskwait_p)
+ g->subcode |= GF_OMP_TASK_TASKWAIT;
+ else
+ g->subcode &= ~GF_OMP_TASK_TASKWAIT;
+}
+
+
/* Return the child function used to hold the body of OMP_TASK GS. */
static inline tree
@@ -5857,6 +5925,60 @@ gimple_omp_teams_set_clauses (gomp_teams
omp_teams_stmt->clauses = clauses;
}
+/* Return the child function used to hold the body of OMP_TEAMS_STMT. */
+
+static inline tree
+gimple_omp_teams_child_fn (const gomp_teams *omp_teams_stmt)
+{
+ return omp_teams_stmt->child_fn;
+}
+
+/* Return a pointer to the child function used to hold the body of
+ OMP_TEAMS_STMT. */
+
+static inline tree *
+gimple_omp_teams_child_fn_ptr (gomp_teams *omp_teams_stmt)
+{
+ return &omp_teams_stmt->child_fn;
+}
+
+
+/* Set CHILD_FN to be the child function for OMP_TEAMS_STMT. */
+
+static inline void
+gimple_omp_teams_set_child_fn (gomp_teams *omp_teams_stmt, tree child_fn)
+{
+ omp_teams_stmt->child_fn = child_fn;
+}
+
+
+/* Return the artificial argument used to send variables and values
+ from the parent to the children threads in OMP_TEAMS_STMT. */
+
+static inline tree
+gimple_omp_teams_data_arg (const gomp_teams *omp_teams_stmt)
+{
+ return omp_teams_stmt->data_arg;
+}
+
+
+/* Return a pointer to the data argument for OMP_TEAMS_STMT. */
+
+static inline tree *
+gimple_omp_teams_data_arg_ptr (gomp_teams *omp_teams_stmt)
+{
+ return &omp_teams_stmt->data_arg;
+}
+
+
+/* Set DATA_ARG to be the data argument for OMP_TEAMS_STMT. */
+
+static inline void
+gimple_omp_teams_set_data_arg (gomp_teams *omp_teams_stmt, tree data_arg)
+{
+ omp_teams_stmt->data_arg = data_arg;
+}
+
/* Return the kernel_phony flag of an OMP_TEAMS_STMT. */
static inline bool
@@ -5876,6 +5998,25 @@ gimple_omp_teams_set_grid_phony (gomp_te
omp_teams_stmt->subcode &= ~GF_OMP_TEAMS_GRID_PHONY;
}
+/* Return the host flag of an OMP_TEAMS_STMT. */
+
+static inline bool
+gimple_omp_teams_host (const gomp_teams *omp_teams_stmt)
+{
+ return (gimple_omp_subcode (omp_teams_stmt) & GF_OMP_TEAMS_HOST) != 0;
+}
+
+/* Set host flag of an OMP_TEAMS_STMT to VALUE. */
+
+static inline void
+gimple_omp_teams_set_host (gomp_teams *omp_teams_stmt, bool value)
+{
+ if (value)
+ omp_teams_stmt->subcode |= GF_OMP_TEAMS_HOST;
+ else
+ omp_teams_stmt->subcode &= ~GF_OMP_TEAMS_HOST;
+}
+
/* Return the clauses associated with OMP_SECTIONS GS. */
static inline tree
@@ -1554,6 +1554,35 @@ dump_gimple_omp_single (pretty_printer *
}
}
+/* Dump a GIMPLE_OMP_TASKGROUP tuple on the pretty_printer BUFFER. */
+
+static void
+dump_gimple_omp_taskgroup (pretty_printer *buffer, gimple *gs,
+ int spc, dump_flags_t flags)
+{
+ if (flags & TDF_RAW)
+ {
+ dump_gimple_fmt (buffer, spc, flags, "%G <%+BODY <%S>%nCLAUSES <", gs,
+ gimple_omp_body (gs));
+ dump_omp_clauses (buffer, gimple_omp_taskgroup_clauses (gs), spc, flags);
+ dump_gimple_fmt (buffer, spc, flags, " >");
+ }
+ else
+ {
+ pp_string (buffer, "#pragma omp taskgroup");
+ dump_omp_clauses (buffer, gimple_omp_taskgroup_clauses (gs), spc, flags);
+ if (!gimple_seq_empty_p (gimple_omp_body (gs)))
+ {
+ newline_and_indent (buffer, spc + 2);
+ pp_left_brace (buffer);
+ pp_newline (buffer);
+ dump_gimple_seq (buffer, gimple_omp_body (gs), spc + 4, flags);
+ newline_and_indent (buffer, spc + 2);
+ pp_right_brace (buffer);
+ }
+ }
+}
+
/* Dump a GIMPLE_OMP_TARGET tuple on the pretty_printer BUFFER. */
static void
@@ -1712,7 +1741,7 @@ dump_gimple_omp_sections (pretty_printer
}
}
-/* Dump a GIMPLE_OMP_{MASTER,TASKGROUP,ORDERED,SECTION} tuple on the
+/* Dump a GIMPLE_OMP_{MASTER,ORDERED,SECTION} tuple on the
pretty_printer BUFFER. */
static void
@@ -2301,6 +2330,8 @@ dump_gimple_omp_task (pretty_printer *bu
gimple_seq body;
if (gimple_omp_task_taskloop_p (gs))
pp_string (buffer, "#pragma omp taskloop");
+ else if (gimple_omp_task_taskwait_p (gs))
+ pp_string (buffer, "#pragma omp taskwait");
else
pp_string (buffer, "#pragma omp task");
dump_omp_clauses (buffer, gimple_omp_task_clauses (gs), spc, flags);
@@ -2353,8 +2384,8 @@ dump_gimple_omp_atomic_load (pretty_prin
else
{
pp_string (buffer, "#pragma omp atomic_load");
- if (gimple_omp_atomic_seq_cst_p (gs))
- pp_string (buffer, " seq_cst");
+ dump_omp_atomic_memory_order (buffer,
+ gimple_omp_atomic_memory_order (gs));
if (gimple_omp_atomic_need_value_p (gs))
pp_string (buffer, " [needed]");
newline_and_indent (buffer, spc + 2);
@@ -2385,9 +2416,10 @@ dump_gimple_omp_atomic_store (pretty_pri
}
else
{
- pp_string (buffer, "#pragma omp atomic_store ");
- if (gimple_omp_atomic_seq_cst_p (gs))
- pp_string (buffer, "seq_cst ");
+ pp_string (buffer, "#pragma omp atomic_store");
+ dump_omp_atomic_memory_order (buffer,
+ gimple_omp_atomic_memory_order (gs));
+ pp_space (buffer);
if (gimple_omp_atomic_need_value_p (gs))
pp_string (buffer, "[needed] ");
pp_left_paren (buffer);
@@ -2569,8 +2601,11 @@ pp_gimple_stmt_1 (pretty_printer *buffer
pp_string (buffer, "GIMPLE_SECTIONS_SWITCH");
break;
- case GIMPLE_OMP_MASTER:
case GIMPLE_OMP_TASKGROUP:
+ dump_gimple_omp_taskgroup (buffer, gs, spc, flags);
+ break;
+
+ case GIMPLE_OMP_MASTER:
case GIMPLE_OMP_SECTION:
case GIMPLE_OMP_GRID_BODY:
dump_gimple_omp_block (buffer, gs, spc, flags);
@@ -105,6 +105,14 @@ 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_NONTEMPORAL = 4194304,
+
GOVD_DATA_SHARE_CLASS = (GOVD_SHARED | GOVD_PRIVATE | GOVD_FIRSTPRIVATE
| GOVD_LASTPRIVATE | GOVD_REDUCTION | GOVD_LINEAR
| GOVD_LOCAL)
@@ -114,34 +122,39 @@ enum gimplify_omp_var_data
enum omp_region_type
{
ORT_WORKSHARE = 0x00,
- ORT_SIMD = 0x01,
-
- ORT_PARALLEL = 0x02,
- ORT_COMBINED_PARALLEL = 0x03,
+ ORT_TASKGROUP = 0x01,
+ ORT_SIMD = 0x04,
- ORT_TASK = 0x04,
- ORT_UNTIED_TASK = 0x05,
+ ORT_PARALLEL = 0x08,
+ ORT_COMBINED_PARALLEL = ORT_PARALLEL | 1,
- ORT_TEAMS = 0x08,
- ORT_COMBINED_TEAMS = 0x09,
+ ORT_TASK = 0x10,
+ ORT_UNTIED_TASK = ORT_TASK | 1,
+ ORT_TASKLOOP = ORT_TASK | 2,
+ ORT_UNTIED_TASKLOOP = ORT_UNTIED_TASK | 2,
+
+ ORT_TEAMS = 0x20,
+ ORT_COMBINED_TEAMS = ORT_TEAMS | 1,
+ ORT_HOST_TEAMS = ORT_TEAMS | 2,
+ ORT_COMBINED_HOST_TEAMS = ORT_COMBINED_TEAMS | 2,
/* Data region. */
- ORT_TARGET_DATA = 0x10,
+ ORT_TARGET_DATA = 0x40,
/* Data region with offloading. */
- ORT_TARGET = 0x20,
- ORT_COMBINED_TARGET = 0x21,
+ ORT_TARGET = 0x80,
+ ORT_COMBINED_TARGET = ORT_TARGET | 1,
/* OpenACC variants. */
- ORT_ACC = 0x40, /* A generic OpenACC region. */
+ ORT_ACC = 0x100, /* A generic OpenACC region. */
ORT_ACC_DATA = ORT_ACC | ORT_TARGET_DATA, /* Data construct. */
ORT_ACC_PARALLEL = ORT_ACC | ORT_TARGET, /* Parallel construct */
- ORT_ACC_KERNELS = ORT_ACC | ORT_TARGET | 0x80, /* Kernels construct. */
- ORT_ACC_HOST_DATA = ORT_ACC | ORT_TARGET_DATA | 0x80, /* Host data. */
+ ORT_ACC_KERNELS = ORT_ACC | ORT_TARGET | 2, /* Kernels construct. */
+ ORT_ACC_HOST_DATA = ORT_ACC | ORT_TARGET_DATA | 2, /* Host data. */
/* Dummy OpenMP region, used to disable expansion of
DECL_VALUE_EXPRs in taskloop pre body. */
- ORT_NONE = 0x100
+ ORT_NONE = 0x200
};
/* Gimplify hashtable helper. */
@@ -176,6 +189,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 +209,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 +433,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;
}
@@ -738,6 +762,7 @@ gimple_add_tmp_var (tree tmp)
struct gimplify_omp_ctx *ctx = gimplify_omp_ctxp;
while (ctx
&& (ctx->region_type == ORT_WORKSHARE
+ || ctx->region_type == ORT_TASKGROUP
|| ctx->region_type == ORT_SIMD
|| ctx->region_type == ORT_ACC))
ctx = ctx->outer_context;
@@ -3150,6 +3175,8 @@ maybe_fold_stmt (gimple_stmt_iterator *g
for (ctx = gimplify_omp_ctxp; ctx; ctx = ctx->outer_context)
if ((ctx->region_type & (ORT_TARGET | ORT_PARALLEL | ORT_TASK)) != 0)
return false;
+ else if ((ctx->region_type & ORT_HOST_TEAMS) == ORT_HOST_TEAMS)
+ return false;
return fold_stmt (gsi);
}
@@ -6641,12 +6668,13 @@ 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);
}
else if (ctx->region_type != ORT_WORKSHARE
+ && ctx->region_type != ORT_TASKGROUP
&& ctx->region_type != ORT_SIMD
&& ctx->region_type != ORT_ACC
&& !(ctx->region_type & ORT_TARGET_DATA))
@@ -6760,7 +6788,7 @@ omp_add_variable (struct gimplify_omp_ct
replacement is private, else FIRSTPRIVATE since we'll need the
address of the original variable either for SHARED, or for the
copy into or out of the context. */
- if (!(flags & GOVD_LOCAL))
+ if (!(flags & GOVD_LOCAL) && ctx->region_type != ORT_TASKGROUP)
{
if (flags & GOVD_MAP)
nflags = GOVD_MAP | GOVD_MAP_TO_ONLY | GOVD_EXPLICIT;
@@ -6962,6 +6990,8 @@ omp_default_clause (struct gimplify_omp_
if (ctx->region_type & ORT_PARALLEL)
rtype = "parallel";
+ else if ((ctx->region_type & ORT_TASKLOOP) == ORT_TASKLOOP)
+ rtype = "taskloop";
else if (ctx->region_type & ORT_TASK)
rtype = "task";
else if (ctx->region_type & ORT_TEAMS)
@@ -7173,11 +7203,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)
{
@@ -7194,18 +7222,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;
@@ -7236,28 +7280,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);
}
@@ -7275,6 +7319,7 @@ omp_notice_variable (struct gimplify_omp
if (n == NULL)
{
if (ctx->region_type == ORT_WORKSHARE
+ || ctx->region_type == ORT_TASKGROUP
|| ctx->region_type == ORT_SIMD
|| ctx->region_type == ORT_ACC
|| (ctx->region_type & ORT_TARGET_DATA) != 0)
@@ -7386,18 +7431,9 @@ omp_is_private (struct gimplify_omp_ctx
else if ((n->value & GOVD_REDUCTION) != 0)
error ("iteration variable %qE should not be reduction",
DECL_NAME (decl));
- else if (simd == 0 && (n->value & GOVD_LINEAR) != 0)
+ else if (simd != 1 && (n->value & GOVD_LINEAR) != 0)
error ("iteration variable %qE should not be linear",
DECL_NAME (decl));
- else if (simd == 1 && (n->value & GOVD_LASTPRIVATE) != 0)
- error ("iteration variable %qE should not be lastprivate",
- DECL_NAME (decl));
- else if (simd && (n->value & GOVD_PRIVATE) != 0)
- error ("iteration variable %qE should not be private",
- DECL_NAME (decl));
- else if (simd == 2 && (n->value & GOVD_LINEAR) != 0)
- error ("iteration variable %qE is predetermined linear",
- DECL_NAME (decl));
}
return (ctx == gimplify_omp_ctxp
|| (ctx->region_type == ORT_COMBINED_PARALLEL
@@ -7405,6 +7441,7 @@ omp_is_private (struct gimplify_omp_ctx
}
if (ctx->region_type != ORT_WORKSHARE
+ && ctx->region_type != ORT_TASKGROUP
&& ctx->region_type != ORT_SIMD
&& ctx->region_type != ORT_ACC)
return false;
@@ -7462,6 +7499,7 @@ omp_check_private (struct gimplify_omp_c
}
}
while (ctx->region_type == ORT_WORKSHARE
+ || ctx->region_type == ORT_TASKGROUP
|| ctx->region_type == ORT_SIMD
|| ctx->region_type == ORT_ACC);
return false;
@@ -7483,6 +7521,452 @@ find_decl_expr (tree *tp, int *walk_subt
return NULL_TREE;
}
+/* If *LIST_P contains any OpenMP depend clauses with iterators,
+ lower all the depend clauses by populating corresponding depend
+ array. Returns 0 if there are no such depend clauses, or
+ 2 if all depend clauses should be removed, 1 otherwise. */
+
+static int
+gimplify_omp_depend (tree *list_p, gimple_seq *pre_p)
+{
+ tree c;
+ gimple *g;
+ size_t n[4] = { 0, 0, 0, 0 };
+ bool unused[4];
+ tree counts[4] = { NULL_TREE, NULL_TREE, NULL_TREE, NULL_TREE };
+ tree last_iter = NULL_TREE, last_count = NULL_TREE;
+ size_t i, j;
+ location_t first_loc = UNKNOWN_LOCATION;
+
+ for (c = *list_p; c; c = OMP_CLAUSE_CHAIN (c))
+ if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_DEPEND)
+ {
+ switch (OMP_CLAUSE_DEPEND_KIND (c))
+ {
+ case OMP_CLAUSE_DEPEND_IN:
+ i = 2;
+ break;
+ case OMP_CLAUSE_DEPEND_OUT:
+ case OMP_CLAUSE_DEPEND_INOUT:
+ i = 0;
+ break;
+ case OMP_CLAUSE_DEPEND_MUTEXINOUTSET:
+ i = 1;
+ break;
+ case OMP_CLAUSE_DEPEND_DEPOBJ:
+ i = 3;
+ break;
+ case OMP_CLAUSE_DEPEND_SOURCE:
+ case OMP_CLAUSE_DEPEND_SINK:
+ continue;
+ default:
+ gcc_unreachable ();
+ }
+ tree t = OMP_CLAUSE_DECL (c);
+ if (first_loc == UNKNOWN_LOCATION)
+ first_loc = OMP_CLAUSE_LOCATION (c);
+ if (TREE_CODE (t) == TREE_LIST
+ && TREE_PURPOSE (t)
+ && TREE_CODE (TREE_PURPOSE (t)) == TREE_VEC)
+ {
+ if (TREE_PURPOSE (t) != last_iter)
+ {
+ tree tcnt = size_one_node;
+ for (tree it = TREE_PURPOSE (t); it; it = TREE_CHAIN (it))
+ {
+ if (gimplify_expr (&TREE_VEC_ELT (it, 1), pre_p, NULL,
+ is_gimple_val, fb_rvalue) == GS_ERROR
+ || gimplify_expr (&TREE_VEC_ELT (it, 2), pre_p, NULL,
+ is_gimple_val, fb_rvalue) == GS_ERROR
+ || gimplify_expr (&TREE_VEC_ELT (it, 3), pre_p, NULL,
+ is_gimple_val, fb_rvalue) == GS_ERROR
+ || (gimplify_expr (&TREE_VEC_ELT (it, 4), pre_p, NULL,
+ is_gimple_val, fb_rvalue)
+ == GS_ERROR))
+ return 2;
+ tree var = TREE_VEC_ELT (it, 0);
+ tree begin = TREE_VEC_ELT (it, 1);
+ tree end = TREE_VEC_ELT (it, 2);
+ tree step = TREE_VEC_ELT (it, 3);
+ tree orig_step = TREE_VEC_ELT (it, 4);
+ tree type = TREE_TYPE (var);
+ tree stype = TREE_TYPE (step);
+ location_t loc = DECL_SOURCE_LOCATION (var);
+ tree endmbegin;
+ /* Compute count for this iterator as
+ orig_step > 0
+ ? (begin < end ? (end - begin + (step - 1)) / step : 0)
+ : (begin > end ? (end - begin + (step + 1)) / step : 0)
+ and compute product of those for the entire depend
+ clause. */
+ if (POINTER_TYPE_P (type))
+ endmbegin = fold_build2_loc (loc, POINTER_DIFF_EXPR,
+ stype, end, begin);
+ else
+ endmbegin = fold_build2_loc (loc, MINUS_EXPR, type,
+ end, begin);
+ tree stepm1 = fold_build2_loc (loc, MINUS_EXPR, stype,
+ step,
+ build_int_cst (stype, 1));
+ tree stepp1 = fold_build2_loc (loc, PLUS_EXPR, stype, step,
+ build_int_cst (stype, 1));
+ tree pos = fold_build2_loc (loc, PLUS_EXPR, stype,
+ unshare_expr (endmbegin),
+ stepm1);
+ pos = fold_build2_loc (loc, TRUNC_DIV_EXPR, stype,
+ pos, step);
+ tree neg = fold_build2_loc (loc, PLUS_EXPR, stype,
+ endmbegin, stepp1);
+ if (TYPE_UNSIGNED (stype))
+ {
+ neg = fold_build1_loc (loc, NEGATE_EXPR, stype, neg);
+ step = fold_build1_loc (loc, NEGATE_EXPR, stype, step);
+ }
+ neg = fold_build2_loc (loc, TRUNC_DIV_EXPR, stype,
+ neg, step);
+ step = NULL_TREE;
+ tree cond = fold_build2_loc (loc, LT_EXPR,
+ boolean_type_node,
+ begin, end);
+ pos = fold_build3_loc (loc, COND_EXPR, stype, cond, pos,
+ build_int_cst (stype, 0));
+ cond = fold_build2_loc (loc, LT_EXPR, boolean_type_node,
+ end, begin);
+ neg = fold_build3_loc (loc, COND_EXPR, stype, cond, neg,
+ build_int_cst (stype, 0));
+ tree osteptype = TREE_TYPE (orig_step);
+ cond = fold_build2_loc (loc, GT_EXPR, boolean_type_node,
+ orig_step,
+ build_int_cst (osteptype, 0));
+ tree cnt = fold_build3_loc (loc, COND_EXPR, stype,
+ cond, pos, neg);
+ cnt = fold_convert_loc (loc, sizetype, cnt);
+ if (gimplify_expr (&cnt, pre_p, NULL, is_gimple_val,
+ fb_rvalue) == GS_ERROR)
+ return 2;
+ tcnt = size_binop_loc (loc, MULT_EXPR, tcnt, cnt);
+ }
+ if (gimplify_expr (&tcnt, pre_p, NULL, is_gimple_val,
+ fb_rvalue) == GS_ERROR)
+ return 2;
+ last_iter = TREE_PURPOSE (t);
+ last_count = tcnt;
+ }
+ if (counts[i] == NULL_TREE)
+ counts[i] = last_count;
+ else
+ counts[i] = size_binop_loc (OMP_CLAUSE_LOCATION (c),
+ PLUS_EXPR, counts[i], last_count);
+ }
+ else
+ n[i]++;
+ }
+ for (i = 0; i < 4; i++)
+ if (counts[i])
+ break;
+ if (i == 4)
+ return 0;
+
+ tree total = size_zero_node;
+ for (i = 0; i < 4; i++)
+ {
+ unused[i] = counts[i] == NULL_TREE && n[i] == 0;
+ if (counts[i] == NULL_TREE)
+ counts[i] = size_zero_node;
+ if (n[i])
+ counts[i] = size_binop (PLUS_EXPR, counts[i], size_int (n[i]));
+ if (gimplify_expr (&counts[i], pre_p, NULL, is_gimple_val,
+ fb_rvalue) == GS_ERROR)
+ return 2;
+ total = size_binop (PLUS_EXPR, total, counts[i]);
+ }
+
+ if (gimplify_expr (&total, pre_p, NULL, is_gimple_val, fb_rvalue)
+ == GS_ERROR)
+ return 2;
+ bool is_old = unused[1] && unused[3];
+ tree totalpx = size_binop (PLUS_EXPR, unshare_expr (total),
+ size_int (is_old ? 1 : 4));
+ tree type = build_array_type (ptr_type_node, build_index_type (totalpx));
+ tree array = create_tmp_var_raw (type);
+ TREE_ADDRESSABLE (array) = 1;
+ if (TREE_CODE (totalpx) != INTEGER_CST)
+ {
+ if (!TYPE_SIZES_GIMPLIFIED (TREE_TYPE (array)))
+ gimplify_type_sizes (TREE_TYPE (array), pre_p);
+ if (gimplify_omp_ctxp)
+ {
+ struct gimplify_omp_ctx *ctx = gimplify_omp_ctxp;
+ while (ctx
+ && (ctx->region_type == ORT_WORKSHARE
+ || ctx->region_type == ORT_TASKGROUP
+ || ctx->region_type == ORT_SIMD
+ || ctx->region_type == ORT_ACC))
+ ctx = ctx->outer_context;
+ if (ctx)
+ omp_add_variable (ctx, array, GOVD_LOCAL | GOVD_SEEN);
+ }
+ gimplify_vla_decl (array, pre_p);
+ }
+ else
+ gimple_add_tmp_var (array);
+ tree r = build4 (ARRAY_REF, ptr_type_node, array, size_int (0), NULL_TREE,
+ NULL_TREE);
+ tree tem;
+ if (!is_old)
+ {
+ tem = build2 (MODIFY_EXPR, void_type_node, r,
+ build_int_cst (ptr_type_node, 0));
+ gimplify_and_add (tem, pre_p);
+ r = build4 (ARRAY_REF, ptr_type_node, array, size_int (1), NULL_TREE,
+ NULL_TREE);
+ }
+ tem = build2 (MODIFY_EXPR, void_type_node, r,
+ fold_convert (ptr_type_node, total));
+ gimplify_and_add (tem, pre_p);
+ for (i = 1; i < (is_old ? 2 : 4); i++)
+ {
+ r = build4 (ARRAY_REF, ptr_type_node, array, size_int (i + !is_old),
+ NULL_TREE, NULL_TREE);
+ tem = build2 (MODIFY_EXPR, void_type_node, r, counts[i - 1]);
+ gimplify_and_add (tem, pre_p);
+ }
+
+ tree cnts[4];
+ for (j = 4; j; j--)
+ if (!unused[j - 1])
+ break;
+ for (i = 0; i < 4; i++)
+ {
+ if (i && (i >= j || unused[i - 1]))
+ {
+ cnts[i] = cnts[i - 1];
+ continue;
+ }
+ cnts[i] = create_tmp_var (sizetype);
+ if (i == 0)
+ g = gimple_build_assign (cnts[i], size_int (is_old ? 2 : 5));
+ else
+ {
+ tree t;
+ if (is_old)
+ t = size_binop (PLUS_EXPR, counts[0], size_int (2));
+ else
+ t = size_binop (PLUS_EXPR, cnts[i - 1], counts[i - 1]);
+ if (gimplify_expr (&t, pre_p, NULL, is_gimple_val, fb_rvalue)
+ == GS_ERROR)
+ return 2;
+ g = gimple_build_assign (cnts[i], t);
+ }
+ gimple_seq_add_stmt (pre_p, g);
+ }
+
+ last_iter = NULL_TREE;
+ tree last_bind = NULL_TREE;
+ tree *last_body = NULL;
+ for (c = *list_p; c; c = OMP_CLAUSE_CHAIN (c))
+ if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_DEPEND)
+ {
+ switch (OMP_CLAUSE_DEPEND_KIND (c))
+ {
+ case OMP_CLAUSE_DEPEND_IN:
+ i = 2;
+ break;
+ case OMP_CLAUSE_DEPEND_OUT:
+ case OMP_CLAUSE_DEPEND_INOUT:
+ i = 0;
+ break;
+ case OMP_CLAUSE_DEPEND_MUTEXINOUTSET:
+ i = 1;
+ break;
+ case OMP_CLAUSE_DEPEND_DEPOBJ:
+ i = 3;
+ break;
+ case OMP_CLAUSE_DEPEND_SOURCE:
+ case OMP_CLAUSE_DEPEND_SINK:
+ continue;
+ default:
+ gcc_unreachable ();
+ }
+ tree t = OMP_CLAUSE_DECL (c);
+ if (TREE_CODE (t) == TREE_LIST
+ && TREE_PURPOSE (t)
+ && TREE_CODE (TREE_PURPOSE (t)) == TREE_VEC)
+ {
+ if (TREE_PURPOSE (t) != last_iter)
+ {
+ if (last_bind)
+ gimplify_and_add (last_bind, pre_p);
+ tree block = TREE_VEC_ELT (TREE_PURPOSE (t), 5);
+ last_bind = build3 (BIND_EXPR, void_type_node,
+ BLOCK_VARS (block), NULL, block);
+ TREE_SIDE_EFFECTS (last_bind) = 1;
+ SET_EXPR_LOCATION (last_bind, OMP_CLAUSE_LOCATION (c));
+ tree *p = &BIND_EXPR_BODY (last_bind);
+ for (tree it = TREE_PURPOSE (t); it; it = TREE_CHAIN (it))
+ {
+ tree var = TREE_VEC_ELT (it, 0);
+ tree begin = TREE_VEC_ELT (it, 1);
+ tree end = TREE_VEC_ELT (it, 2);
+ tree step = TREE_VEC_ELT (it, 3);
+ tree orig_step = TREE_VEC_ELT (it, 4);
+ tree type = TREE_TYPE (var);
+ location_t loc = DECL_SOURCE_LOCATION (var);
+ /* Emit:
+ var = begin;
+ goto cond_label;
+ beg_label:
+ ...
+ var = var + step;
+ cond_label:
+ if (orig_step > 0) {
+ if (var < end) goto beg_label;
+ } else {
+ if (var > end) goto beg_label;
+ }
+ for each iterator, with inner iterators added to
+ the ... above. */
+ tree beg_label = create_artificial_label (loc);
+ tree cond_label = NULL_TREE;
+ tem = build2_loc (loc, MODIFY_EXPR, void_type_node,
+ var, begin);
+ append_to_statement_list_force (tem, p);
+ tem = build_and_jump (&cond_label);
+ append_to_statement_list_force (tem, p);
+ tem = build1 (LABEL_EXPR, void_type_node, beg_label);
+ append_to_statement_list (tem, p);
+ tree bind = build3 (BIND_EXPR, void_type_node, NULL_TREE,
+ NULL_TREE, NULL_TREE);
+ TREE_SIDE_EFFECTS (bind) = 1;
+ SET_EXPR_LOCATION (bind, loc);
+ append_to_statement_list_force (bind, p);
+ if (POINTER_TYPE_P (type))
+ tem = build2_loc (loc, POINTER_PLUS_EXPR, type,
+ var, fold_convert_loc (loc, sizetype,
+ step));
+ else
+ tem = build2_loc (loc, PLUS_EXPR, type, var, step);
+ tem = build2_loc (loc, MODIFY_EXPR, void_type_node,
+ var, tem);
+ append_to_statement_list_force (tem, p);
+ tem = build1 (LABEL_EXPR, void_type_node, cond_label);
+ append_to_statement_list (tem, p);
+ tree cond = fold_build2_loc (loc, LT_EXPR,
+ boolean_type_node,
+ var, end);
+ tree pos
+ = fold_build3_loc (loc, COND_EXPR, void_type_node,
+ cond, build_and_jump (&beg_label),
+ void_node);
+ cond = fold_build2_loc (loc, GT_EXPR, boolean_type_node,
+ var, end);
+ tree neg
+ = fold_build3_loc (loc, COND_EXPR, void_type_node,
+ cond, build_and_jump (&beg_label),
+ void_node);
+ tree osteptype = TREE_TYPE (orig_step);
+ cond = fold_build2_loc (loc, GT_EXPR, boolean_type_node,
+ orig_step,
+ build_int_cst (osteptype, 0));
+ tem = fold_build3_loc (loc, COND_EXPR, void_type_node,
+ cond, pos, neg);
+ append_to_statement_list_force (tem, p);
+ p = &BIND_EXPR_BODY (bind);
+ }
+ last_body = p;
+ }
+ last_iter = TREE_PURPOSE (t);
+ if (TREE_CODE (TREE_VALUE (t)) == COMPOUND_EXPR)
+ {
+ append_to_statement_list (TREE_OPERAND (TREE_VALUE (t),
+ 0), last_body);
+ TREE_VALUE (t) = TREE_OPERAND (TREE_VALUE (t), 1);
+ }
+ if (error_operand_p (TREE_VALUE (t)))
+ return 2;
+ TREE_VALUE (t) = build_fold_addr_expr (TREE_VALUE (t));
+ r = build4 (ARRAY_REF, ptr_type_node, array, cnts[i],
+ NULL_TREE, NULL_TREE);
+ tem = build2_loc (OMP_CLAUSE_LOCATION (c), MODIFY_EXPR,
+ void_type_node, r, TREE_VALUE (t));
+ append_to_statement_list_force (tem, last_body);
+ tem = build2_loc (OMP_CLAUSE_LOCATION (c), MODIFY_EXPR,
+ void_type_node, cnts[i],
+ size_binop (PLUS_EXPR, cnts[i], size_int (1)));
+ append_to_statement_list_force (tem, last_body);
+ TREE_VALUE (t) = null_pointer_node;
+ }
+ else
+ {
+ if (last_bind)
+ {
+ gimplify_and_add (last_bind, pre_p);
+ last_bind = NULL_TREE;
+ }
+ if (TREE_CODE (OMP_CLAUSE_DECL (c)) == COMPOUND_EXPR)
+ {
+ gimplify_expr (&TREE_OPERAND (OMP_CLAUSE_DECL (c), 0), pre_p,
+ NULL, is_gimple_val, fb_rvalue);
+ OMP_CLAUSE_DECL (c) = TREE_OPERAND (OMP_CLAUSE_DECL (c), 1);
+ }
+ if (error_operand_p (OMP_CLAUSE_DECL (c)))
+ return 2;
+ OMP_CLAUSE_DECL (c) = build_fold_addr_expr (OMP_CLAUSE_DECL (c));
+ if (gimplify_expr (&OMP_CLAUSE_DECL (c), pre_p, NULL,
+ is_gimple_val, fb_rvalue) == GS_ERROR)
+ return 2;
+ r = build4 (ARRAY_REF, ptr_type_node, array, cnts[i],
+ NULL_TREE, NULL_TREE);
+ tem = build2 (MODIFY_EXPR, void_type_node, r, OMP_CLAUSE_DECL (c));
+ gimplify_and_add (tem, pre_p);
+ g = gimple_build_assign (cnts[i], size_binop (PLUS_EXPR, cnts[i],
+ size_int (1)));
+ gimple_seq_add_stmt (pre_p, g);
+ }
+ }
+ if (last_bind)
+ gimplify_and_add (last_bind, pre_p);
+ tree cond = boolean_false_node;
+ if (is_old)
+ {
+ if (!unused[0])
+ cond = build2_loc (first_loc, NE_EXPR, boolean_type_node, cnts[0],
+ size_binop_loc (first_loc, PLUS_EXPR, counts[0],
+ size_int (2)));
+ if (!unused[2])
+ cond = build2_loc (first_loc, TRUTH_OR_EXPR, boolean_type_node, cond,
+ build2_loc (first_loc, NE_EXPR, boolean_type_node,
+ cnts[2],
+ size_binop_loc (first_loc, PLUS_EXPR,
+ totalpx,
+ size_int (1))));
+ }
+ else
+ {
+ tree prev = size_int (5);
+ for (i = 0; i < 4; i++)
+ {
+ if (unused[i])
+ continue;
+ prev = size_binop_loc (first_loc, PLUS_EXPR, counts[i], prev);
+ cond = build2_loc (first_loc, TRUTH_OR_EXPR, boolean_type_node, cond,
+ build2_loc (first_loc, NE_EXPR, boolean_type_node,
+ cnts[i], unshare_expr (prev)));
+ }
+ }
+ tem = build3_loc (first_loc, COND_EXPR, void_type_node, cond,
+ build_call_expr_loc (first_loc,
+ builtin_decl_explicit (BUILT_IN_TRAP),
+ 0), void_node);
+ gimplify_and_add (tem, pre_p);
+ c = build_omp_clause (UNKNOWN_LOCATION, OMP_CLAUSE_DEPEND);
+ OMP_CLAUSE_DEPEND_KIND (c) = OMP_CLAUSE_DEPEND_LAST;
+ OMP_CLAUSE_DECL (c) = build_fold_addr_expr (array);
+ OMP_CLAUSE_CHAIN (c) = *list_p;
+ *list_p = c;
+ return 1;
+}
+
/* Scan the OMP clauses in *LIST_P, installing mappings into a new
and previous omp contexts. */
@@ -7495,14 +7979,16 @@ gimplify_scan_omp_clauses (tree *list_p,
tree c;
hash_map<tree, tree> *struct_map_to_clause = NULL;
tree *prev_list_p = NULL;
+ int handled_depend_iterators = -1;
+ int nowait = -1;
ctx = new_omp_context (region_type);
outer_ctx = ctx->outer_context;
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)
@@ -7548,16 +8034,43 @@ gimplify_scan_omp_clauses (tree *list_p,
check_non_private = "firstprivate";
goto do_add;
case OMP_CLAUSE_LASTPRIVATE:
+ if (OMP_CLAUSE_LASTPRIVATE_CONDITIONAL (c))
+ switch (code)
+ {
+ case OMP_DISTRIBUTE:
+ error_at (OMP_CLAUSE_LOCATION (c),
+ "conditional %<lastprivate%> clause on "
+ "%<distribute%> construct");
+ OMP_CLAUSE_LASTPRIVATE_CONDITIONAL (c) = 0;
+ break;
+ case OMP_TASKLOOP:
+ error_at (OMP_CLAUSE_LOCATION (c),
+ "conditional %<lastprivate%> clause on "
+ "%<taskloop%> construct");
+ OMP_CLAUSE_LASTPRIVATE_CONDITIONAL (c) = 0;
+ break;
+ default:
+ break;
+ }
flags = GOVD_LASTPRIVATE | GOVD_SEEN | GOVD_EXPLICIT;
check_non_private = "lastprivate";
decl = OMP_CLAUSE_DECL (c);
if (error_operand_p (decl))
goto do_add;
- else if (outer_ctx
- && (outer_ctx->region_type == ORT_COMBINED_PARALLEL
- || outer_ctx->region_type == ORT_COMBINED_TEAMS)
- && splay_tree_lookup (outer_ctx->variables,
- (splay_tree_key) decl) == NULL)
+ if (OMP_CLAUSE_LASTPRIVATE_CONDITIONAL (c)
+ && !lang_hooks.decls.omp_scalar_p (decl))
+ {
+ error_at (OMP_CLAUSE_LOCATION (c),
+ "non-scalar variable %qD in conditional "
+ "%<lastprivate%> clause", decl);
+ OMP_CLAUSE_LASTPRIVATE_CONDITIONAL (c) = 0;
+ }
+ if (outer_ctx
+ && (outer_ctx->region_type == ORT_COMBINED_PARALLEL
+ || ((outer_ctx->region_type & ORT_COMBINED_TEAMS)
+ == ORT_COMBINED_TEAMS))
+ && splay_tree_lookup (outer_ctx->variables,
+ (splay_tree_key) decl) == NULL)
{
omp_add_variable (outer_ctx, decl, GOVD_SHARED | GOVD_SEEN);
if (outer_ctx->outer_context)
@@ -7603,7 +8116,8 @@ gimplify_scan_omp_clauses (tree *list_p,
GOVD_LASTPRIVATE | GOVD_SEEN);
octx = octx->outer_context;
if (octx
- && octx->region_type == ORT_COMBINED_TEAMS
+ && ((octx->region_type & ORT_COMBINED_TEAMS)
+ == ORT_COMBINED_TEAMS)
&& (splay_tree_lookup (octx->variables,
(splay_tree_key) decl)
== NULL))
@@ -7622,10 +8136,40 @@ gimplify_scan_omp_clauses (tree *list_p,
}
goto do_add;
case OMP_CLAUSE_REDUCTION:
+ if (OMP_CLAUSE_REDUCTION_TASK (c))
+ {
+ if (region_type == ORT_WORKSHARE)
+ {
+ if (nowait == -1)
+ nowait = omp_find_clause (*list_p,
+ OMP_CLAUSE_NOWAIT) != NULL_TREE;
+ if (nowait
+ && (outer_ctx == NULL
+ || outer_ctx->region_type != ORT_COMBINED_PARALLEL))
+ {
+ error_at (OMP_CLAUSE_LOCATION (c),
+ "%<task%> reduction modifier on a construct "
+ "with a %<nowait%> clause");
+ OMP_CLAUSE_REDUCTION_TASK (c) = 0;
+ }
+ }
+ else if ((region_type & ORT_PARALLEL) != ORT_PARALLEL)
+ {
+ error_at (OMP_CLAUSE_LOCATION (c),
+ "invalid %<task%> reduction modifier on construct "
+ "other than %<parallel%>, %<for%> or %<sections%>");
+ OMP_CLAUSE_REDUCTION_TASK (c) = 0;
+ }
+ }
+ /* FALLTHRU */
+ case OMP_CLAUSE_IN_REDUCTION:
+ case OMP_CLAUSE_TASK_REDUCTION:
flags = GOVD_REDUCTION | GOVD_SEEN | GOVD_EXPLICIT;
/* OpenACC permits reductions on private variables. */
- if (!(region_type & ORT_ACC))
- check_non_private = "reduction";
+ if (!(region_type & ORT_ACC)
+ /* taskgroup is actually not a worksharing region. */
+ && code != OMP_TASKGROUP)
+ check_non_private = omp_clause_code_name[OMP_CLAUSE_CODE (c)];
decl = OMP_CLAUSE_DECL (c);
if (TREE_CODE (decl) == MEM_REF)
{
@@ -7746,7 +8290,8 @@ gimplify_scan_omp_clauses (tree *list_p,
&& octx == outer_ctx)
flags = GOVD_SEEN | GOVD_SHARED;
else if (octx
- && octx->region_type == ORT_COMBINED_TEAMS)
+ && ((octx->region_type & ORT_COMBINED_TEAMS)
+ == ORT_COMBINED_TEAMS))
flags = GOVD_SEEN | GOVD_SHARED;
else if (octx
&& octx->region_type == ORT_COMBINED_TARGET)
@@ -8229,6 +8774,14 @@ gimplify_scan_omp_clauses (tree *list_p,
}
else if (OMP_CLAUSE_DEPEND_KIND (c) == OMP_CLAUSE_DEPEND_SOURCE)
break;
+ if (handled_depend_iterators == -1)
+ handled_depend_iterators = gimplify_omp_depend (list_p, pre_p);
+ if (handled_depend_iterators)
+ {
+ if (handled_depend_iterators == 2)
+ remove = true;
+ break;
+ }
if (TREE_CODE (OMP_CLAUSE_DECL (c)) == COMPOUND_EXPR)
{
gimplify_expr (&TREE_OPERAND (OMP_CLAUSE_DECL (c), 0), pre_p,
@@ -8311,7 +8864,9 @@ gimplify_scan_omp_clauses (tree *list_p,
&& OMP_CLAUSE_MAP_KIND (c) == GOMP_MAP_FIRSTPRIVATE_POINTER)
flags |= GOVD_MAP_0LEN_ARRAY;
omp_add_variable (ctx, decl, flags);
- if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_REDUCTION
+ if ((OMP_CLAUSE_CODE (c) == OMP_CLAUSE_REDUCTION
+ || OMP_CLAUSE_CODE (c) == OMP_CLAUSE_IN_REDUCTION
+ || OMP_CLAUSE_CODE (c) == OMP_CLAUSE_TASK_REDUCTION)
&& OMP_CLAUSE_REDUCTION_PLACEHOLDER (c))
{
omp_add_variable (ctx, OMP_CLAUSE_REDUCTION_PLACEHOLDER (c),
@@ -8423,6 +8978,31 @@ gimplify_scan_omp_clauses (tree *list_p,
" or private in outer context", DECL_NAME (decl));
}
do_notice:
+ if ((region_type & ORT_TASKLOOP) == ORT_TASKLOOP
+ && outer_ctx
+ && outer_ctx->region_type == ORT_COMBINED_PARALLEL
+ && (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_REDUCTION
+ || OMP_CLAUSE_CODE (c) == OMP_CLAUSE_FIRSTPRIVATE
+ || OMP_CLAUSE_CODE (c) == OMP_CLAUSE_LASTPRIVATE))
+ {
+ splay_tree_node on
+ = splay_tree_lookup (outer_ctx->variables,
+ (splay_tree_key)decl);
+ if (on == NULL || (on->value & GOVD_DATA_SHARE_CLASS) == 0)
+ {
+ if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_REDUCTION
+ && TREE_CODE (OMP_CLAUSE_DECL (c)) == MEM_REF
+ && (TREE_CODE (TREE_TYPE (decl)) == POINTER_TYPE
+ || (TREE_CODE (TREE_TYPE (decl)) == REFERENCE_TYPE
+ && (TREE_CODE (TREE_TYPE (TREE_TYPE (decl)))
+ == POINTER_TYPE))))
+ omp_firstprivatize_variable (outer_ctx, decl);
+ else
+ omp_add_variable (outer_ctx, decl,
+ GOVD_SEEN | GOVD_SHARED);
+ omp_notice_variable (outer_ctx, decl, true);
+ }
+ }
if (outer_ctx)
omp_notice_variable (outer_ctx, decl, true);
if (check_non_private
@@ -8453,7 +9033,9 @@ gimplify_scan_omp_clauses (tree *list_p,
for (int i = 0; i < 2; i++)
switch (i ? OMP_CLAUSE_IF_MODIFIER (c) : code)
{
+ case VOID_CST: p[i] = "cancel"; break;
case OMP_PARALLEL: p[i] = "parallel"; break;
+ case OMP_SIMD: p[i] = "simd"; break;
case OMP_TASK: p[i] = "task"; break;
case OMP_TASKLOOP: p[i] = "taskloop"; break;
case OMP_TARGET_DATA: p[i] = "target data"; break;
@@ -8508,6 +9090,9 @@ gimplify_scan_omp_clauses (tree *list_p,
break;
case OMP_CLAUSE_NOWAIT:
+ nowait = 1;
+ break;
+
case OMP_CLAUSE_ORDERED:
case OMP_CLAUSE_UNTIED:
case OMP_CLAUSE_COLLAPSE:
@@ -8527,7 +9112,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:
@@ -8548,6 +9195,16 @@ gimplify_scan_omp_clauses (tree *list_p,
omp_add_variable (ctx, decl, GOVD_ALIGNED);
break;
+ case OMP_CLAUSE_NONTEMPORAL:
+ decl = OMP_CLAUSE_DECL (c);
+ if (error_operand_p (decl))
+ {
+ remove = true;
+ break;
+ }
+ omp_add_variable (ctx, decl, GOVD_NONTEMPORAL);
+ break;
+
case OMP_CLAUSE_DEFAULT:
ctx->default_kind = OMP_CLAUSE_DEFAULT_KIND (c);
break;
@@ -8773,7 +9430,7 @@ gimplify_adjust_omp_clauses_1 (splay_tre
}
else if (flags & GOVD_LASTPRIVATE)
code = OMP_CLAUSE_LASTPRIVATE;
- else if (flags & GOVD_ALIGNED)
+ else if (flags & (GOVD_ALIGNED | GOVD_NONTEMPORAL))
return 0;
else
gcc_unreachable ();
@@ -8827,7 +9484,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;
@@ -8838,6 +9497,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;
@@ -9062,6 +9727,12 @@ gimplify_adjust_omp_clauses (gimple_seq
}
break;
+ case OMP_CLAUSE_NONTEMPORAL:
+ decl = OMP_CLAUSE_DECL (c);
+ n = splay_tree_lookup (ctx->variables, (splay_tree_key) decl);
+ remove = n == NULL || !(n->value & GOVD_SEEN);
+ break;
+
case OMP_CLAUSE_MAP:
if (code == OMP_TARGET_EXIT_DATA
&& OMP_CLAUSE_MAP_KIND (c) == GOMP_MAP_ALWAYS_POINTER)
@@ -9229,6 +9900,8 @@ gimplify_adjust_omp_clauses (gimple_seq
break;
case OMP_CLAUSE_REDUCTION:
+ case OMP_CLAUSE_IN_REDUCTION:
+ case OMP_CLAUSE_TASK_REDUCTION:
decl = OMP_CLAUSE_DECL (c);
/* OpenACC reductions need a present_or_copy data clause.
Add one if necessary. Emit error when the reduction is private. */
@@ -9506,18 +10179,32 @@ gimplify_omp_task (tree *expr_p, gimple_
gimple *g;
gimple_seq body = NULL;
+ if (OMP_TASK_BODY (expr) == NULL_TREE)
+ for (tree c = OMP_TASK_CLAUSES (expr); c; c = OMP_CLAUSE_CHAIN (c))
+ if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_DEPEND
+ && OMP_CLAUSE_DEPEND_KIND (c) == OMP_CLAUSE_DEPEND_MUTEXINOUTSET)
+ {
+ error_at (OMP_CLAUSE_LOCATION (c),
+ "%<mutexinoutset%> kind in %<depend%> clause on a "
+ "%<taskwait%> construct");
+ break;
+ }
+
gimplify_scan_omp_clauses (&OMP_TASK_CLAUSES (expr), pre_p,
omp_find_clause (OMP_TASK_CLAUSES (expr),
OMP_CLAUSE_UNTIED)
? ORT_UNTIED_TASK : ORT_TASK, OMP_TASK);
- push_gimplify_context ();
+ if (OMP_TASK_BODY (expr))
+ {
+ push_gimplify_context ();
- g = gimplify_and_return_first (OMP_TASK_BODY (expr), &body);
- if (gimple_code (g) == GIMPLE_BIND)
- pop_gimplify_context (g);
- else
- pop_gimplify_context (NULL);
+ g = gimplify_and_return_first (OMP_TASK_BODY (expr), &body);
+ if (gimple_code (g) == GIMPLE_BIND)
+ pop_gimplify_context (g);
+ else
+ pop_gimplify_context (NULL);
+ }
gimplify_adjust_omp_clauses (pre_p, body, &OMP_TASK_CLAUSES (expr),
OMP_TASK);
@@ -9526,6 +10213,8 @@ gimplify_omp_task (tree *expr_p, gimple_
OMP_TASK_CLAUSES (expr),
NULL_TREE, NULL_TREE,
NULL_TREE, NULL_TREE, NULL_TREE);
+ if (OMP_TASK_BODY (expr) == NULL_TREE)
+ gimple_omp_task_set_taskwait_p (g, true);
gimplify_seq_add_stmt (pre_p, g);
*expr_p = NULL_TREE;
}
@@ -9655,7 +10344,9 @@ gimplify_omp_for (tree *expr_p, gimple_s
for (i = 0; i < TREE_VEC_LENGTH (OMP_FOR_INIT (inner_for_stmt)); i++)
if (OMP_FOR_ORIG_DECLS (inner_for_stmt)
&& TREE_CODE (TREE_VEC_ELT (OMP_FOR_ORIG_DECLS (inner_for_stmt),
- i)) == TREE_LIST)
+ i)) == TREE_LIST
+ && TREE_PURPOSE (TREE_VEC_ELT (OMP_FOR_ORIG_DECLS (inner_for_stmt),
+ i)))
{
tree orig = TREE_VEC_ELT (OMP_FOR_ORIG_DECLS (inner_for_stmt), i);
/* Class iterators aren't allowed on OMP_SIMD, so the only
@@ -9709,6 +10400,43 @@ gimplify_omp_for (tree *expr_p, gimple_s
OMP_CLAUSE_CHAIN (c) = OMP_PARALLEL_CLAUSES (*data[1]);
OMP_PARALLEL_CLAUSES (*data[1]) = c;
}
+ /* Similarly, take care of C++ range for temporaries, those should
+ be firstprivate on OMP_PARALLEL if any. */
+ if (data[1])
+ for (i = 0; i < TREE_VEC_LENGTH (OMP_FOR_INIT (inner_for_stmt)); i++)
+ if (OMP_FOR_ORIG_DECLS (inner_for_stmt)
+ && TREE_CODE (TREE_VEC_ELT (OMP_FOR_ORIG_DECLS (inner_for_stmt),
+ i)) == TREE_LIST
+ && TREE_CHAIN (TREE_VEC_ELT (OMP_FOR_ORIG_DECLS (inner_for_stmt),
+ i)))
+ {
+ tree orig
+ = TREE_VEC_ELT (OMP_FOR_ORIG_DECLS (inner_for_stmt), i);
+ tree v = TREE_CHAIN (orig);
+ tree c = build_omp_clause (UNKNOWN_LOCATION,
+ OMP_CLAUSE_FIRSTPRIVATE);
+ /* First add firstprivate clause for the __for_end artificial
+ decl. */
+ OMP_CLAUSE_DECL (c) = TREE_VEC_ELT (v, 1);
+ if (TREE_CODE (TREE_TYPE (OMP_CLAUSE_DECL (c)))
+ == REFERENCE_TYPE)
+ OMP_CLAUSE_FIRSTPRIVATE_NO_REFERENCE (c) = 1;
+ OMP_CLAUSE_CHAIN (c) = OMP_PARALLEL_CLAUSES (*data[1]);
+ OMP_PARALLEL_CLAUSES (*data[1]) = c;
+ if (TREE_VEC_ELT (v, 0))
+ {
+ /* And now the same for __for_range artificial decl if it
+ exists. */
+ c = build_omp_clause (UNKNOWN_LOCATION,
+ OMP_CLAUSE_FIRSTPRIVATE);
+ OMP_CLAUSE_DECL (c) = TREE_VEC_ELT (v, 0);
+ if (TREE_CODE (TREE_TYPE (OMP_CLAUSE_DECL (c)))
+ == REFERENCE_TYPE)
+ OMP_CLAUSE_FIRSTPRIVATE_NO_REFERENCE (c) = 1;
+ OMP_CLAUSE_CHAIN (c) = OMP_PARALLEL_CLAUSES (*data[1]);
+ OMP_PARALLEL_CLAUSES (*data[1]) = c;
+ }
+ }
}
switch (TREE_CODE (for_stmt))
@@ -9721,9 +10449,9 @@ gimplify_omp_for (tree *expr_p, gimple_s
break;
case OMP_TASKLOOP:
if (omp_find_clause (OMP_FOR_CLAUSES (for_stmt), OMP_CLAUSE_UNTIED))
- ort = ORT_UNTIED_TASK;
+ ort = ORT_UNTIED_TASKLOOP;
else
- ort = ORT_TASK;
+ ort = ORT_TASKLOOP;
break;
case OMP_SIMD:
ort = ORT_SIMD;
@@ -9935,7 +10663,11 @@ gimplify_omp_for (tree *expr_p, gimple_s
{
tree orig_decl = TREE_VEC_ELT (OMP_FOR_ORIG_DECLS (for_stmt), i);
if (TREE_CODE (orig_decl) == TREE_LIST)
- orig_decl = TREE_PURPOSE (orig_decl);
+ {
+ orig_decl = TREE_PURPOSE (orig_decl);
+ if (!orig_decl)
+ orig_decl = decl;
+ }
gimplify_omp_ctxp->loop_iter_var.quick_push (orig_decl);
}
else
@@ -10027,7 +10759,8 @@ gimplify_omp_for (tree *expr_p, gimple_s
else if (omp_check_private (outer, decl, false))
outer = NULL;
}
- else if (((outer->region_type & ORT_TASK) != 0)
+ else if (((outer->region_type & ORT_TASKLOOP)
+ == ORT_TASKLOOP)
&& outer->combined_loop
&& !omp_check_private (gimplify_omp_ctxp,
decl, false))
@@ -10066,8 +10799,12 @@ gimplify_omp_for (tree *expr_p, gimple_s
outer = NULL;
}
if (outer && outer->outer_context
- && (outer->outer_context->region_type
- == ORT_COMBINED_TEAMS))
+ && ((outer->outer_context->region_type
+ & ORT_COMBINED_TEAMS) == ORT_COMBINED_TEAMS
+ || (((outer->region_type & ORT_TASKLOOP)
+ == ORT_TASKLOOP)
+ && (outer->outer_context->region_type
+ == ORT_COMBINED_PARALLEL))))
{
outer = outer->outer_context;
n = splay_tree_lookup (outer->variables,
@@ -10114,7 +10851,8 @@ gimplify_omp_for (tree *expr_p, gimple_s
else if (omp_check_private (outer, decl, false))
outer = NULL;
}
- else if (((outer->region_type & ORT_TASK) != 0)
+ else if (((outer->region_type & ORT_TASKLOOP)
+ == ORT_TASKLOOP)
&& outer->combined_loop
&& !omp_check_private (gimplify_omp_ctxp,
decl, false))
@@ -10153,8 +10891,12 @@ gimplify_omp_for (tree *expr_p, gimple_s
outer = NULL;
}
if (outer && outer->outer_context
- && (outer->outer_context->region_type
- == ORT_COMBINED_TEAMS))
+ && ((outer->outer_context->region_type
+ & ORT_COMBINED_TEAMS) == ORT_COMBINED_TEAMS
+ || (((outer->region_type & ORT_TASKLOOP)
+ == ORT_TASKLOOP)
+ && (outer->outer_context->region_type
+ == ORT_COMBINED_PARALLEL))))
{
outer = outer->outer_context;
n = splay_tree_lookup (outer->variables,
@@ -10496,6 +11238,8 @@ gimplify_omp_for (tree *expr_p, gimple_s
case OMP_CLAUSE_FINAL:
case OMP_CLAUSE_MERGEABLE:
case OMP_CLAUSE_PRIORITY:
+ case OMP_CLAUSE_REDUCTION:
+ case OMP_CLAUSE_IN_REDUCTION:
*gtask_clauses_ptr = c;
gtask_clauses_ptr = &OMP_CLAUSE_CHAIN (c);
break;
@@ -10676,7 +11420,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;
}
@@ -10848,6 +11592,12 @@ gimplify_omp_workshare (tree *expr_p, gi
break;
case OMP_TEAMS:
ort = OMP_TEAMS_COMBINED (expr) ? ORT_COMBINED_TEAMS : ORT_TEAMS;
+ if (gimplify_omp_ctxp == NULL
+ || (gimplify_omp_ctxp->region_type == ORT_TARGET
+ && gimplify_omp_ctxp->outer_context == NULL
+ && lookup_attribute ("omp declare target",
+ DECL_ATTRIBUTES (current_function_decl))))
+ ort = (enum omp_region_type) (ort | ORT_HOST_TEAMS);
break;
case OACC_HOST_DATA:
ort = ORT_ACC_HOST_DATA;
@@ -10859,7 +11609,8 @@ gimplify_omp_workshare (tree *expr_p, gi
TREE_CODE (expr));
if (TREE_CODE (expr) == OMP_TARGET)
optimize_target_teams (expr, pre_p);
- if ((ort & (ORT_TARGET | ORT_TARGET_DATA)) != 0)
+ if ((ort & (ORT_TARGET | ORT_TARGET_DATA)) != 0
+ || (ort & ORT_HOST_TEAMS) == ORT_HOST_TEAMS)
{
push_gimplify_context ();
gimple *g = gimplify_and_return_first (OMP_BODY (expr), &body);
@@ -10930,6 +11681,8 @@ gimplify_omp_workshare (tree *expr_p, gi
break;
case OMP_TEAMS:
stmt = gimple_build_omp_teams (body, OMP_CLAUSES (expr));
+ if ((ort & ORT_HOST_TEAMS) == ORT_HOST_TEAMS)
+ gimple_omp_teams_set_host (as_a <gomp_teams *> (stmt), true);
break;
default:
gcc_unreachable ();
@@ -11165,7 +11918,8 @@ gimplify_omp_atomic (tree *expr_p, gimpl
!= GS_ALL_DONE)
return GS_ERROR;
- loadstmt = gimple_build_omp_atomic_load (tmp_load, addr);
+ loadstmt = gimple_build_omp_atomic_load (tmp_load, addr,
+ OMP_ATOMIC_MEMORY_ORDER (*expr_p));
gimplify_seq_add_stmt (pre_p, loadstmt);
if (rhs && gimplify_expr (&rhs, pre_p, NULL, is_gimple_val, fb_rvalue)
!= GS_ALL_DONE)
@@ -11173,13 +11927,9 @@ gimplify_omp_atomic (tree *expr_p, gimpl
if (TREE_CODE (*expr_p) == OMP_ATOMIC_READ)
rhs = tmp_load;
- storestmt = gimple_build_omp_atomic_store (rhs);
+ storestmt
+ = gimple_build_omp_atomic_store (rhs, OMP_ATOMIC_MEMORY_ORDER (*expr_p));
gimplify_seq_add_stmt (pre_p, storestmt);
- if (OMP_ATOMIC_SEQ_CST (*expr_p))
- {
- gimple_omp_atomic_set_seq_cst (loadstmt);
- gimple_omp_atomic_set_seq_cst (storestmt);
- }
switch (TREE_CODE (*expr_p))
{
case OMP_ATOMIC_READ:
@@ -12126,7 +12876,6 @@ gimplify_expr (tree *expr_p, gimple_seq
case OMP_SECTION:
case OMP_MASTER:
- case OMP_TASKGROUP:
case OMP_ORDERED:
case OMP_CRITICAL:
{
@@ -12142,19 +12891,6 @@ gimplify_expr (tree *expr_p, gimple_seq
case OMP_MASTER:
g = gimple_build_omp_master (body);
break;
- case OMP_TASKGROUP:
- {
- gimple_seq cleanup = NULL;
- tree fn
- = builtin_decl_explicit (BUILT_IN_GOMP_TASKGROUP_END);
- g = gimple_build_call (fn, 0);
- gimple_seq_add_stmt (&cleanup, g);
- g = gimple_build_try (body, cleanup, GIMPLE_TRY_FINALLY);
- body = NULL;
- gimple_seq_add_stmt (&body, g);
- g = gimple_build_omp_taskgroup (body);
- }
- break;
case OMP_ORDERED:
g = gimplify_omp_ordered (*expr_p, body);
break;
@@ -12174,6 +12910,28 @@ gimplify_expr (tree *expr_p, gimple_seq
gimplify_seq_add_stmt (pre_p, g);
ret = GS_ALL_DONE;
break;
+ }
+
+ case OMP_TASKGROUP:
+ {
+ gimple_seq body = NULL;
+
+ tree *pclauses = &OMP_TASKGROUP_CLAUSES (*expr_p);
+ gimplify_scan_omp_clauses (pclauses, pre_p, ORT_TASKGROUP,
+ OMP_TASKGROUP);
+ gimplify_adjust_omp_clauses (pre_p, NULL, pclauses, OMP_TASKGROUP);
+ gimplify_and_add (OMP_BODY (*expr_p), &body);
+ gimple_seq cleanup = NULL;
+ tree fn = builtin_decl_explicit (BUILT_IN_GOMP_TASKGROUP_END);
+ gimple *g = gimple_build_call (fn, 0);
+ gimple_seq_add_stmt (&cleanup, g);
+ g = gimple_build_try (body, cleanup, GIMPLE_TRY_FINALLY);
+ body = NULL;
+ gimple_seq_add_stmt (&body, g);
+ g = gimple_build_omp_taskgroup (body, *pclauses);
+ gimplify_seq_add_stmt (pre_p, g);
+ ret = GS_ALL_DONE;
+ break;
}
case OMP_ATOMIC:
@@ -1356,6 +1356,8 @@ hash_tree (struct streamer_tree_cache_d
val = OMP_CLAUSE_PROC_BIND_KIND (t);
break;
case OMP_CLAUSE_REDUCTION:
+ case OMP_CLAUSE_TASK_REDUCTION:
+ case OMP_CLAUSE_IN_REDUCTION:
val = OMP_CLAUSE_REDUCTION_CODE (t);
break;
default:
@@ -2578,6 +2578,7 @@ GTFILES = $(CPPLIB_H) $(srcdir)/input.h
$(srcdir)/internal-fn.h \
$(srcdir)/hsa-common.c \
$(srcdir)/calls.c \
+ $(srcdir)/omp-general.h \
@all_gtfiles@
# Compute the list of GT header files from the corresponding C sources,
@@ -75,6 +75,8 @@ DEF_GOMP_BUILTIN (BUILT_IN_GOMP_BARRIER_
BT_FN_BOOL, ATTR_NOTHROW_LEAF_LIST)
DEF_GOMP_BUILTIN (BUILT_IN_GOMP_TASKWAIT, "GOMP_taskwait",
BT_FN_VOID, ATTR_NOTHROW_LEAF_LIST)
+DEF_GOMP_BUILTIN (BUILT_IN_GOMP_TASKWAIT_DEPEND, "GOMP_taskwait_depend",
+ BT_FN_VOID_PTR, ATTR_NOTHROW_LEAF_LIST)
DEF_GOMP_BUILTIN (BUILT_IN_GOMP_TASKYIELD, "GOMP_taskyield",
BT_FN_VOID, ATTR_NOTHROW_LEAF_LIST)
DEF_GOMP_BUILTIN (BUILT_IN_GOMP_TASKGROUP_START, "GOMP_taskgroup_start",
@@ -122,6 +124,14 @@ DEF_GOMP_BUILTIN (BUILT_IN_GOMP_LOOP_NON
"GOMP_loop_nonmonotonic_guided_start",
BT_FN_BOOL_LONG_LONG_LONG_LONG_LONGPTR_LONGPTR,
ATTR_NOTHROW_LEAF_LIST)
+DEF_GOMP_BUILTIN (BUILT_IN_GOMP_LOOP_NONMONOTONIC_RUNTIME_START,
+ "GOMP_loop_nonmonotonic_runtime_start",
+ BT_FN_BOOL_LONG_LONG_LONG_LONG_LONGPTR_LONGPTR,
+ ATTR_NOTHROW_LEAF_LIST)
+DEF_GOMP_BUILTIN (BUILT_IN_GOMP_LOOP_MAYBE_NONMONOTONIC_RUNTIME_START,
+ "GOMP_loop_maybe_nonmonotonic_runtime_start",
+ BT_FN_BOOL_LONG_LONG_LONG_LONG_LONGPTR_LONGPTR,
+ ATTR_NOTHROW_LEAF_LIST)
DEF_GOMP_BUILTIN (BUILT_IN_GOMP_LOOP_ORDERED_STATIC_START,
"GOMP_loop_ordered_static_start",
BT_FN_BOOL_LONG_LONG_LONG_LONG_LONGPTR_LONGPTR,
@@ -154,6 +164,18 @@ DEF_GOMP_BUILTIN (BUILT_IN_GOMP_LOOP_DOA
"GOMP_loop_doacross_runtime_start",
BT_FN_BOOL_UINT_LONGPTR_LONGPTR_LONGPTR,
ATTR_NOTHROW_LEAF_LIST)
+DEF_GOMP_BUILTIN (BUILT_IN_GOMP_LOOP_START,
+ "GOMP_loop_start",
+ BT_FN_BOOL_LONG_LONG_LONG_LONG_LONG_LONGPTR_LONGPTR_PTR_PTR,
+ ATTR_NOTHROW_LEAF_LIST)
+DEF_GOMP_BUILTIN (BUILT_IN_GOMP_LOOP_ORDERED_START,
+ "GOMP_loop_ordered_start",
+ BT_FN_BOOL_LONG_LONG_LONG_LONG_LONG_LONGPTR_LONGPTR_PTR_PTR,
+ ATTR_NOTHROW_LEAF_LIST)
+DEF_GOMP_BUILTIN (BUILT_IN_GOMP_LOOP_DOACROSS_START,
+ "GOMP_loop_doacross_start",
+ BT_FN_BOOL_UINT_LONGPTR_LONG_LONG_LONGPTR_LONGPTR_PTR_PTR,
+ ATTR_NOTHROW_LEAF_LIST)
DEF_GOMP_BUILTIN (BUILT_IN_GOMP_LOOP_STATIC_NEXT, "GOMP_loop_static_next",
BT_FN_BOOL_LONGPTR_LONGPTR, ATTR_NOTHROW_LEAF_LIST)
DEF_GOMP_BUILTIN (BUILT_IN_GOMP_LOOP_DYNAMIC_NEXT, "GOMP_loop_dynamic_next",
@@ -168,6 +190,12 @@ DEF_GOMP_BUILTIN (BUILT_IN_GOMP_LOOP_NON
DEF_GOMP_BUILTIN (BUILT_IN_GOMP_LOOP_NONMONOTONIC_GUIDED_NEXT,
"GOMP_loop_nonmonotonic_guided_next",
BT_FN_BOOL_LONGPTR_LONGPTR, ATTR_NOTHROW_LEAF_LIST)
+DEF_GOMP_BUILTIN (BUILT_IN_GOMP_LOOP_NONMONOTONIC_RUNTIME_NEXT,
+ "GOMP_loop_nonmonotonic_runtime_next",
+ BT_FN_BOOL_LONGPTR_LONGPTR, ATTR_NOTHROW_LEAF_LIST)
+DEF_GOMP_BUILTIN (BUILT_IN_GOMP_LOOP_MAYBE_NONMONOTONIC_RUNTIME_NEXT,
+ "GOMP_loop_maybe_nonmonotonic_runtime_next",
+ BT_FN_BOOL_LONGPTR_LONGPTR, ATTR_NOTHROW_LEAF_LIST)
DEF_GOMP_BUILTIN (BUILT_IN_GOMP_LOOP_ORDERED_STATIC_NEXT,
"GOMP_loop_ordered_static_next",
BT_FN_BOOL_LONGPTR_LONGPTR, ATTR_NOTHROW_LEAF_LIST)
@@ -204,6 +232,14 @@ DEF_GOMP_BUILTIN (BUILT_IN_GOMP_LOOP_ULL
"GOMP_loop_ull_nonmonotonic_guided_start",
BT_FN_BOOL_BOOL_ULL_ULL_ULL_ULL_ULLPTR_ULLPTR,
ATTR_NOTHROW_LEAF_LIST)
+DEF_GOMP_BUILTIN (BUILT_IN_GOMP_LOOP_ULL_NONMONOTONIC_RUNTIME_START,
+ "GOMP_loop_ull_nonmonotonic_runtime_start",
+ BT_FN_BOOL_BOOL_ULL_ULL_ULL_ULL_ULLPTR_ULLPTR,
+ ATTR_NOTHROW_LEAF_LIST)
+DEF_GOMP_BUILTIN (BUILT_IN_GOMP_LOOP_ULL_MAYBE_NONMONOTONIC_RUNTIME_START,
+ "GOMP_loop_ull_maybe_nonmonotonic_runtime_start",
+ BT_FN_BOOL_BOOL_ULL_ULL_ULL_ULL_ULLPTR_ULLPTR,
+ ATTR_NOTHROW_LEAF_LIST)
DEF_GOMP_BUILTIN (BUILT_IN_GOMP_LOOP_ULL_ORDERED_STATIC_START,
"GOMP_loop_ull_ordered_static_start",
BT_FN_BOOL_BOOL_ULL_ULL_ULL_ULL_ULLPTR_ULLPTR,
@@ -236,6 +272,18 @@ DEF_GOMP_BUILTIN (BUILT_IN_GOMP_LOOP_ULL
"GOMP_loop_ull_doacross_runtime_start",
BT_FN_BOOL_UINT_ULLPTR_ULLPTR_ULLPTR,
ATTR_NOTHROW_LEAF_LIST)
+DEF_GOMP_BUILTIN (BUILT_IN_GOMP_LOOP_ULL_START,
+ "GOMP_loop_ull_start",
+ BT_FN_BOOL_BOOL_ULL_ULL_ULL_LONG_ULL_ULLPTR_ULLPTR_PTR_PTR,
+ ATTR_NOTHROW_LEAF_LIST)
+DEF_GOMP_BUILTIN (BUILT_IN_GOMP_LOOP_ULL_ORDERED_START,
+ "GOMP_loop_ull_ordered_start",
+ BT_FN_BOOL_BOOL_ULL_ULL_ULL_LONG_ULL_ULLPTR_ULLPTR_PTR_PTR,
+ ATTR_NOTHROW_LEAF_LIST)
+DEF_GOMP_BUILTIN (BUILT_IN_GOMP_LOOP_ULL_DOACROSS_START,
+ "GOMP_loop_ull_doacross_start",
+ BT_FN_BOOL_UINT_ULLPTR_LONG_ULL_ULLPTR_ULLPTR_PTR_PTR,
+ ATTR_NOTHROW_LEAF_LIST)
DEF_GOMP_BUILTIN (BUILT_IN_GOMP_LOOP_ULL_STATIC_NEXT,
"GOMP_loop_ull_static_next",
BT_FN_BOOL_ULONGLONGPTR_ULONGLONGPTR, ATTR_NOTHROW_LEAF_LIST)
@@ -254,6 +302,12 @@ DEF_GOMP_BUILTIN (BUILT_IN_GOMP_LOOP_ULL
DEF_GOMP_BUILTIN (BUILT_IN_GOMP_LOOP_ULL_NONMONOTONIC_GUIDED_NEXT,
"GOMP_loop_ull_nonmonotonic_guided_next",
BT_FN_BOOL_ULONGLONGPTR_ULONGLONGPTR, ATTR_NOTHROW_LEAF_LIST)
+DEF_GOMP_BUILTIN (BUILT_IN_GOMP_LOOP_ULL_NONMONOTONIC_RUNTIME_NEXT,
+ "GOMP_loop_ull_nonmonotonic_runtime_next",
+ BT_FN_BOOL_ULONGLONGPTR_ULONGLONGPTR, ATTR_NOTHROW_LEAF_LIST)
+DEF_GOMP_BUILTIN (BUILT_IN_GOMP_LOOP_ULL_MAYBE_NONMONOTONIC_RUNTIME_NEXT,
+ "GOMP_loop_ull_maybe_nonmonotonic_runtime_next",
+ BT_FN_BOOL_ULONGLONGPTR_ULONGLONGPTR, ATTR_NOTHROW_LEAF_LIST)
DEF_GOMP_BUILTIN (BUILT_IN_GOMP_LOOP_ULL_ORDERED_STATIC_NEXT,
"GOMP_loop_ull_ordered_static_next",
BT_FN_BOOL_ULONGLONGPTR_ULONGLONGPTR, ATTR_NOTHROW_LEAF_LIST)
@@ -293,6 +347,14 @@ DEF_GOMP_BUILTIN (BUILT_IN_GOMP_PARALLEL
"GOMP_parallel_loop_nonmonotonic_guided",
BT_FN_VOID_OMPFN_PTR_UINT_LONG_LONG_LONG_LONG_UINT,
ATTR_NOTHROW_LIST)
+DEF_GOMP_BUILTIN (BUILT_IN_GOMP_PARALLEL_LOOP_NONMONOTONIC_RUNTIME,
+ "GOMP_parallel_loop_nonmonotonic_runtime",
+ BT_FN_VOID_OMPFN_PTR_UINT_LONG_LONG_LONG_LONG_UINT,
+ ATTR_NOTHROW_LIST)
+DEF_GOMP_BUILTIN (BUILT_IN_GOMP_PARALLEL_LOOP_MAYBE_NONMONOTONIC_RUNTIME,
+ "GOMP_parallel_loop_maybe_nonmonotonic_runtime",
+ BT_FN_VOID_OMPFN_PTR_UINT_LONG_LONG_LONG_LONG_UINT,
+ ATTR_NOTHROW_LIST)
DEF_GOMP_BUILTIN (BUILT_IN_GOMP_LOOP_END, "GOMP_loop_end",
BT_FN_VOID, ATTR_NOTHROW_LEAF_LIST)
DEF_GOMP_BUILTIN (BUILT_IN_GOMP_LOOP_END_CANCEL, "GOMP_loop_end_cancel",
@@ -313,6 +375,9 @@ DEF_GOMP_BUILTIN (BUILT_IN_GOMP_DOACROSS
BT_FN_VOID_ULL_VAR, ATTR_NOTHROW_LEAF_LIST)
DEF_GOMP_BUILTIN (BUILT_IN_GOMP_PARALLEL, "GOMP_parallel",
BT_FN_VOID_OMPFN_PTR_UINT_UINT, ATTR_NOTHROW_LIST)
+DEF_GOMP_BUILTIN (BUILT_IN_GOMP_PARALLEL_REDUCTIONS,
+ "GOMP_parallel_reductions",
+ BT_FN_UINT_OMPFN_PTR_UINT_UINT, ATTR_NOTHROW_LIST)
DEF_GOMP_BUILTIN (BUILT_IN_GOMP_TASK, "GOMP_task",
BT_FN_VOID_OMPFN_PTR_OMPCPYFN_LONG_LONG_BOOL_UINT_PTR_INT,
ATTR_NOTHROW_LIST)
@@ -324,6 +389,8 @@ DEF_GOMP_BUILTIN (BUILT_IN_GOMP_TASKLOOP
ATTR_NOTHROW_LIST)
DEF_GOMP_BUILTIN (BUILT_IN_GOMP_SECTIONS_START, "GOMP_sections_start",
BT_FN_UINT_UINT, ATTR_NOTHROW_LEAF_LIST)
+DEF_GOMP_BUILTIN (BUILT_IN_GOMP_SECTIONS2_START, "GOMP_sections2_start",
+ BT_FN_UINT_UINT_PTR_PTR, ATTR_NOTHROW_LEAF_LIST)
DEF_GOMP_BUILTIN (BUILT_IN_GOMP_SECTIONS_NEXT, "GOMP_sections_next",
BT_FN_UINT, ATTR_NOTHROW_LEAF_LIST)
DEF_GOMP_BUILTIN (BUILT_IN_GOMP_PARALLEL_SECTIONS,
@@ -363,5 +430,19 @@ DEF_GOMP_BUILTIN (BUILT_IN_GOMP_TARGET_E
BT_FN_VOID_INT_SIZE_PTR_PTR_PTR_UINT_PTR, ATTR_NOTHROW_LIST)
DEF_GOMP_BUILTIN (BUILT_IN_GOMP_TEAMS, "GOMP_teams",
BT_FN_VOID_UINT_UINT, ATTR_NOTHROW_LIST)
+DEF_GOMP_BUILTIN (BUILT_IN_GOMP_TEAMS_REG, "GOMP_teams_reg",
+ BT_FN_VOID_OMPFN_PTR_UINT_UINT_UINT, ATTR_NOTHROW_LIST)
+DEF_GOMP_BUILTIN (BUILT_IN_GOMP_TASKGROUP_REDUCTION_REGISTER,
+ "GOMP_taskgroup_reduction_register",
+ BT_FN_VOID_PTR, ATTR_NOTHROW_LEAF_LIST)
+DEF_GOMP_BUILTIN (BUILT_IN_GOMP_TASKGROUP_REDUCTION_UNREGISTER,
+ "GOMP_taskgroup_reduction_unregister",
+ BT_FN_VOID_PTR, ATTR_NOTHROW_LEAF_LIST)
+DEF_GOMP_BUILTIN (BUILT_IN_GOMP_TASK_REDUCTION_REMAP,
+ "GOMP_task_reduction_remap",
+ BT_FN_VOID_SIZE_SIZE_PTR, ATTR_NOTHROW_LEAF_LIST)
+DEF_GOMP_BUILTIN (BUILT_IN_GOMP_WORKSHARE_TASK_REDUCTION_UNREGISTER,
+ "GOMP_workshare_task_reduction_unregister",
+ BT_FN_VOID_BOOL, ATTR_NOTHROW_LEAF_LIST)
DEF_GOACC_BUILTIN (BUILT_IN_GOACC_DECLARE, "GOACC_declare",
BT_FN_VOID_INT_SIZE_PTR_PTR_PTR, ATTR_NOTHROW_LIST)
@@ -174,6 +174,8 @@ workshare_safe_to_combine_p (basic_block
return true;
gcc_assert (gimple_code (ws_stmt) == GIMPLE_OMP_FOR);
+ if (gimple_omp_for_kind (ws_stmt) != GF_OMP_FOR_KIND_FOR)
+ return false;
omp_extract_for_data (as_a <gomp_for *> (ws_stmt), &fd, NULL);
@@ -202,7 +204,7 @@ workshare_safe_to_combine_p (basic_block
static tree
omp_adjust_chunk_size (tree chunk_size, bool simd_schedule)
{
- if (!simd_schedule)
+ if (!simd_schedule || integer_zerop (chunk_size))
return chunk_size;
poly_uint64 vf = omp_max_vf ();
@@ -310,6 +312,13 @@ determine_parallel_type (struct omp_regi
ws_entry_bb = region->inner->entry;
ws_exit_bb = region->inner->exit;
+ /* Give up for task reductions on the parallel, while it is implementable,
+ adding another big set of APIs or slowing down the normal paths is
+ not acceptable. */
+ tree pclauses = gimple_omp_parallel_clauses (last_stmt (par_entry_bb));
+ if (omp_find_clause (pclauses, OMP_CLAUSE__REDUCTEMP_))
+ return;
+
if (single_succ (par_entry_bb) == ws_entry_bb
&& single_succ (ws_exit_bb) == par_exit_bb
&& workshare_safe_to_combine_p (ws_entry_bb)
@@ -336,13 +345,14 @@ determine_parallel_type (struct omp_regi
if (c == NULL
|| ((OMP_CLAUSE_SCHEDULE_KIND (c) & OMP_CLAUSE_SCHEDULE_MASK)
== OMP_CLAUSE_SCHEDULE_STATIC)
- || omp_find_clause (clauses, OMP_CLAUSE_ORDERED))
- {
- region->is_combined_parallel = false;
- region->inner->is_combined_parallel = false;
- return;
- }
+ || omp_find_clause (clauses, OMP_CLAUSE_ORDERED)
+ || omp_find_clause (clauses, OMP_CLAUSE__REDUCTEMP_))
+ return;
}
+ else if (region->inner->type == GIMPLE_OMP_SECTIONS
+ && omp_find_clause (gimple_omp_sections_clauses (ws_stmt),
+ OMP_CLAUSE__REDUCTEMP_))
+ return;
region->is_combined_parallel = true;
region->inner->is_combined_parallel = true;
@@ -534,7 +544,7 @@ adjust_context_and_scope (tree entry_blo
}
}
-/* Build the function calls to GOMP_parallel_start etc to actually
+/* Build the function calls to GOMP_parallel etc to actually
generate the parallel operation. REGION is the parallel region
being expanded. BB is the block where to insert the code. WS_ARGS
will be set if this is a call to a combined parallel+workshare
@@ -559,7 +569,10 @@ expand_parallel_call (struct omp_region
/* Determine what flavor of GOMP_parallel we will be
emitting. */
start_ix = BUILT_IN_GOMP_PARALLEL;
- if (is_combined_parallel (region))
+ tree rtmp = omp_find_clause (clauses, OMP_CLAUSE__REDUCTEMP_);
+ if (rtmp)
+ start_ix = BUILT_IN_GOMP_PARALLEL_REDUCTIONS;
+ else if (is_combined_parallel (region))
{
switch (region->inner->type)
{
@@ -568,12 +581,19 @@ expand_parallel_call (struct omp_region
switch (region->inner->sched_kind)
{
case OMP_CLAUSE_SCHEDULE_RUNTIME:
- start_ix2 = 3;
+ if ((region->inner->sched_modifiers
+ & OMP_CLAUSE_SCHEDULE_NONMONOTONIC) != 0)
+ start_ix2 = 6;
+ else if ((region->inner->sched_modifiers
+ & OMP_CLAUSE_SCHEDULE_MONOTONIC) == 0)
+ start_ix2 = 7;
+ else
+ start_ix2 = 3;
break;
case OMP_CLAUSE_SCHEDULE_DYNAMIC:
case OMP_CLAUSE_SCHEDULE_GUIDED:
- if (region->inner->sched_modifiers
- & OMP_CLAUSE_SCHEDULE_NONMONOTONIC)
+ if ((region->inner->sched_modifiers
+ & OMP_CLAUSE_SCHEDULE_MONOTONIC) == 0)
{
start_ix2 = 3 + region->inner->sched_kind;
break;
@@ -716,6 +736,13 @@ expand_parallel_call (struct omp_region
t = build_call_expr_loc_vec (UNKNOWN_LOCATION,
builtin_decl_explicit (start_ix), args);
+ if (rtmp)
+ {
+ tree type = TREE_TYPE (OMP_CLAUSE_DECL (rtmp));
+ t = build2 (MODIFY_EXPR, type, OMP_CLAUSE_DECL (rtmp),
+ fold_convert (type,
+ fold_convert (pointer_sized_int_node, t)));
+ }
force_gimple_operand_gsi (&gsi, t, true, NULL_TREE,
false, GSI_CONTINUE_LINKING);
@@ -792,6 +819,8 @@ expand_task_call (struct omp_region *reg
if (omp_find_clause (tclauses, OMP_CLAUSE_NOGROUP))
iflags |= GOMP_TASK_FLAG_NOGROUP;
ull = fd.iter_type == long_long_unsigned_type_node;
+ if (omp_find_clause (clauses, OMP_CLAUSE_REDUCTION))
+ iflags |= GOMP_TASK_FLAG_REDUCTION;
}
else if (priority)
iflags |= GOMP_TASK_FLAG_PRIORITY;
@@ -866,6 +895,82 @@ expand_task_call (struct omp_region *reg
false, GSI_CONTINUE_LINKING);
}
+/* Build the function call to GOMP_taskwait_depend to actually
+ generate the taskwait operation. BB is the block where to insert the
+ code. */
+
+static void
+expand_taskwait_call (basic_block bb, gomp_task *entry_stmt)
+{
+ tree clauses = gimple_omp_task_clauses (entry_stmt);
+ tree depend = omp_find_clause (clauses, OMP_CLAUSE_DEPEND);
+ if (depend == NULL_TREE)
+ return;
+
+ depend = OMP_CLAUSE_DECL (depend);
+
+ gimple_stmt_iterator gsi = gsi_last_nondebug_bb (bb);
+ tree t
+ = build_call_expr (builtin_decl_explicit (BUILT_IN_GOMP_TASKWAIT_DEPEND),
+ 1, depend);
+
+ force_gimple_operand_gsi (&gsi, t, true, NULL_TREE,
+ false, GSI_CONTINUE_LINKING);
+}
+
+/* Build the function call to GOMP_teams_reg to actually
+ generate the host teams operation. REGION is the teams region
+ being expanded. BB is the block where to insert the code. */
+
+static void
+expand_teams_call (basic_block bb, gomp_teams *entry_stmt)
+{
+ tree clauses = gimple_omp_teams_clauses (entry_stmt);
+ tree num_teams = omp_find_clause (clauses, OMP_CLAUSE_NUM_TEAMS);
+ if (num_teams == NULL_TREE)
+ num_teams = build_int_cst (unsigned_type_node, 0);
+ else
+ {
+ num_teams = OMP_CLAUSE_NUM_TEAMS_EXPR (num_teams);
+ num_teams = fold_convert (unsigned_type_node, num_teams);
+ }
+ tree thread_limit = omp_find_clause (clauses, OMP_CLAUSE_THREAD_LIMIT);
+ if (thread_limit == NULL_TREE)
+ thread_limit = build_int_cst (unsigned_type_node, 0);
+ else
+ {
+ thread_limit = OMP_CLAUSE_THREAD_LIMIT_EXPR (thread_limit);
+ thread_limit = fold_convert (unsigned_type_node, thread_limit);
+ }
+
+ gimple_stmt_iterator gsi = gsi_last_nondebug_bb (bb);
+ tree t = gimple_omp_teams_data_arg (entry_stmt), t1;
+ if (t == NULL)
+ t1 = null_pointer_node;
+ else
+ t1 = build_fold_addr_expr (t);
+ tree child_fndecl = gimple_omp_teams_child_fn (entry_stmt);
+ tree t2 = build_fold_addr_expr (child_fndecl);
+
+ adjust_context_and_scope (gimple_block (entry_stmt), child_fndecl);
+
+ vec<tree, va_gc> *args;
+ vec_alloc (args, 5);
+ args->quick_push (t2);
+ args->quick_push (t1);
+ args->quick_push (num_teams);
+ args->quick_push (thread_limit);
+ /* For future extensibility. */
+ args->quick_push (build_zero_cst (unsigned_type_node));
+
+ t = build_call_expr_loc_vec (UNKNOWN_LOCATION,
+ builtin_decl_explicit (BUILT_IN_GOMP_TEAMS_REG),
+ args);
+
+ force_gimple_operand_gsi (&gsi, t, true, NULL_TREE,
+ false, GSI_CONTINUE_LINKING);
+}
+
/* Chain all the DECLs in LIST by their TREE_CHAIN fields. */
static tree
@@ -1112,6 +1217,17 @@ expand_omp_taskreg (struct omp_region *r
vec<tree, va_gc> *ws_args;
entry_stmt = last_stmt (region->entry);
+ if (gimple_code (entry_stmt) == GIMPLE_OMP_TASK
+ && gimple_omp_task_taskwait_p (entry_stmt))
+ {
+ new_bb = region->entry;
+ gsi = gsi_last_nondebug_bb (region->entry);
+ gcc_assert (gimple_code (gsi_stmt (gsi)) == GIMPLE_OMP_TASK);
+ gsi_remove (&gsi, true);
+ expand_taskwait_call (new_bb, as_a <gomp_task *> (entry_stmt));
+ return;
+ }
+
child_fn = gimple_omp_taskreg_child_fn (entry_stmt);
child_cfun = DECL_STRUCT_FUNCTION (child_fn);
@@ -1137,7 +1253,8 @@ expand_omp_taskreg (struct omp_region *r
gsi = gsi_last_nondebug_bb (entry_bb);
gcc_assert (gimple_code (gsi_stmt (gsi)) == GIMPLE_OMP_PARALLEL
- || gimple_code (gsi_stmt (gsi)) == GIMPLE_OMP_TASK);
+ || gimple_code (gsi_stmt (gsi)) == GIMPLE_OMP_TASK
+ || gimple_code (gsi_stmt (gsi)) == GIMPLE_OMP_TEAMS);
gsi_remove (&gsi, true);
new_bb = entry_bb;
@@ -1190,8 +1307,8 @@ expand_omp_taskreg (struct omp_region *r
effectively doing a STRIP_NOPS. */
if (TREE_CODE (arg) == ADDR_EXPR
- && TREE_OPERAND (arg, 0)
- == gimple_omp_taskreg_data_arg (entry_stmt))
+ && (TREE_OPERAND (arg, 0)
+ == gimple_omp_taskreg_data_arg (entry_stmt)))
{
parcopy_stmt = stmt;
break;
@@ -1251,12 +1368,13 @@ expand_omp_taskreg (struct omp_region *r
gsi = gsi_last_nondebug_bb (entry_bb);
stmt = gsi_stmt (gsi);
gcc_assert (stmt && (gimple_code (stmt) == GIMPLE_OMP_PARALLEL
- || gimple_code (stmt) == GIMPLE_OMP_TASK));
+ || gimple_code (stmt) == GIMPLE_OMP_TASK
+ || gimple_code (stmt) == GIMPLE_OMP_TEAMS));
e = split_block (entry_bb, stmt);
gsi_remove (&gsi, true);
entry_bb = e->dest;
edge e2 = NULL;
- if (gimple_code (entry_stmt) == GIMPLE_OMP_PARALLEL)
+ if (gimple_code (entry_stmt) != GIMPLE_OMP_TASK)
single_succ_edge (entry_bb)->flags = EDGE_FALLTHRU;
else
{
@@ -1382,6 +1500,8 @@ expand_omp_taskreg (struct omp_region *r
if (gimple_code (entry_stmt) == GIMPLE_OMP_PARALLEL)
expand_parallel_call (region, new_bb,
as_a <gomp_parallel *> (entry_stmt), ws_args);
+ else if (gimple_code (entry_stmt) == GIMPLE_OMP_TEAMS)
+ expand_teams_call (new_bb, as_a <gomp_teams *> (entry_stmt));
else
expand_task_call (region, new_bb, as_a <gomp_task *> (entry_stmt));
if (gimple_in_ssa_p (cfun))
@@ -2499,6 +2619,7 @@ expand_omp_for_generic (struct omp_regio
struct omp_for_data *fd,
enum built_in_function start_fn,
enum built_in_function next_fn,
+ tree sched_arg,
gimple *inner_stmt)
{
tree type, istart0, iend0, iend;
@@ -2546,6 +2667,30 @@ expand_omp_for_generic (struct omp_regio
&& omp_find_clause (gimple_omp_for_clauses (gsi_stmt (gsi)),
OMP_CLAUSE_LASTPRIVATE))
ordered_lastprivate = false;
+ tree reductions = NULL_TREE;
+ tree mem = NULL_TREE;
+ if (sched_arg)
+ {
+ if (fd->have_reductemp)
+ {
+ tree c = omp_find_clause (gimple_omp_for_clauses (gsi_stmt (gsi)),
+ OMP_CLAUSE__REDUCTEMP_);
+ reductions = OMP_CLAUSE_DECL (c);
+ gcc_assert (TREE_CODE (reductions) == SSA_NAME);
+ gimple *g = SSA_NAME_DEF_STMT (reductions);
+ reductions = gimple_assign_rhs1 (g);
+ OMP_CLAUSE_DECL (c) = reductions;
+ entry_bb = gimple_bb (g);
+ edge e = split_block (entry_bb, g);
+ if (region->entry == entry_bb)
+ region->entry = e->dest;
+ gsi = gsi_last_bb (entry_bb);
+ }
+ else
+ reductions = null_pointer_node;
+ /* For now. */
+ mem = null_pointer_node;
+ }
if (fd->collapse > 1 || fd->ordered)
{
int first_zero_iter1 = -1, first_zero_iter2 = -1;
@@ -2732,7 +2877,18 @@ expand_omp_for_generic (struct omp_regio
{
t = fold_convert (fd->iter_type, fd->chunk_size);
t = omp_adjust_chunk_size (t, fd->simd_schedule);
- if (fd->ordered)
+ if (sched_arg)
+ {
+ if (fd->ordered)
+ t = build_call_expr (builtin_decl_explicit (start_fn),
+ 8, t0, t1, sched_arg, t, t3, t4,
+ reductions, mem);
+ else
+ t = build_call_expr (builtin_decl_explicit (start_fn),
+ 9, t0, t1, t2, sched_arg, t, t3, t4,
+ reductions, mem);
+ }
+ else if (fd->ordered)
t = build_call_expr (builtin_decl_explicit (start_fn),
5, t0, t1, t, t3, t4);
else
@@ -2765,7 +2921,11 @@ expand_omp_for_generic (struct omp_regio
tree bfn_decl = builtin_decl_explicit (start_fn);
t = fold_convert (fd->iter_type, fd->chunk_size);
t = omp_adjust_chunk_size (t, fd->simd_schedule);
- t = build_call_expr (bfn_decl, 7, t5, t0, t1, t2, t, t3, t4);
+ if (sched_arg)
+ t = build_call_expr (bfn_decl, 10, t5, t0, t1, t2, sched_arg,
+ t, t3, t4, reductions, mem);
+ else
+ t = build_call_expr (bfn_decl, 7, t5, t0, t1, t2, t, t3, t4);
}
else
t = build_call_expr (builtin_decl_explicit (start_fn),
@@ -2784,6 +2944,17 @@ expand_omp_for_generic (struct omp_regio
gsi_insert_before (&gsi, gimple_build_assign (arr, clobber),
GSI_SAME_STMT);
}
+ if (fd->have_reductemp)
+ {
+ gimple *g = gsi_stmt (gsi);
+ gsi_remove (&gsi, true);
+ release_ssa_name (gimple_assign_lhs (g));
+
+ entry_bb = region->entry;
+ gsi = gsi_last_nondebug_bb (entry_bb);
+
+ gcc_assert (gimple_code (gsi_stmt (gsi)) == GIMPLE_OMP_FOR);
+ }
gsi_insert_after (&gsi, gimple_build_cond_empty (t), GSI_SAME_STMT);
/* Remove the GIMPLE_OMP_FOR statement. */
@@ -3082,9 +3253,6 @@ expand_omp_for_generic (struct omp_regio
else
t = builtin_decl_explicit (BUILT_IN_GOMP_LOOP_END);
gcall *call_stmt = gimple_build_call (t, 0);
- if (gimple_omp_return_lhs (gsi_stmt (gsi)))
- gimple_call_set_lhs (call_stmt, gimple_omp_return_lhs (gsi_stmt (gsi)));
- gsi_insert_after (&gsi, call_stmt, GSI_SAME_STMT);
if (fd->ordered)
{
tree arr = counts[fd->ordered];
@@ -3093,6 +3261,17 @@ expand_omp_for_generic (struct omp_regio
gsi_insert_after (&gsi, gimple_build_assign (arr, clobber),
GSI_SAME_STMT);
}
+ if (gimple_omp_return_lhs (gsi_stmt (gsi)))
+ {
+ gimple_call_set_lhs (call_stmt, gimple_omp_return_lhs (gsi_stmt (gsi)));
+ if (fd->have_reductemp)
+ {
+ gimple *g = gimple_build_assign (reductions, NOP_EXPR,
+ gimple_call_lhs (call_stmt));
+ gsi_insert_after (&gsi, g, GSI_SAME_STMT);
+ }
+ }
+ gsi_insert_after (&gsi, call_stmt, GSI_SAME_STMT);
gsi_remove (&gsi, true);
/* Connect the new blocks. */
@@ -3275,6 +3454,7 @@ expand_omp_for_static_nochunk (struct om
bool broken_loop = region->cont == NULL;
tree *counts = NULL;
tree n1, n2, step;
+ tree reductions = NULL_TREE;
itype = type = TREE_TYPE (fd->loop.v);
if (POINTER_TYPE_P (type))
@@ -3358,6 +3538,29 @@ expand_omp_for_static_nochunk (struct om
gsi = gsi_last_bb (entry_bb);
}
+ if (fd->have_reductemp)
+ {
+ tree t1 = build_int_cst (long_integer_type_node, 0);
+ tree t2 = build_int_cst (long_integer_type_node, 1);
+ tree t3 = build_int_cstu (long_integer_type_node,
+ (HOST_WIDE_INT_1U << 31) + 1);
+ tree clauses = gimple_omp_for_clauses (fd->for_stmt);
+ clauses = omp_find_clause (clauses, OMP_CLAUSE__REDUCTEMP_);
+ reductions = OMP_CLAUSE_DECL (clauses);
+ gcc_assert (TREE_CODE (reductions) == SSA_NAME);
+ gimple *g = SSA_NAME_DEF_STMT (reductions);
+ reductions = gimple_assign_rhs1 (g);
+ OMP_CLAUSE_DECL (clauses) = reductions;
+ gimple_stmt_iterator gsi2 = gsi_for_stmt (g);
+ tree t
+ = build_call_expr (builtin_decl_explicit (BUILT_IN_GOMP_LOOP_START),
+ 9, t1, t2, t2, t3, t1, null_pointer_node,
+ null_pointer_node, reductions, null_pointer_node);
+ force_gimple_operand_gsi (&gsi2, t, true, NULL_TREE,
+ true, GSI_SAME_STMT);
+ gsi_remove (&gsi2, true);
+ release_ssa_name (gimple_assign_lhs (g));
+ }
switch (gimple_omp_for_kind (fd->for_stmt))
{
case GF_OMP_FOR_KIND_FOR:
@@ -3628,7 +3831,25 @@ expand_omp_for_static_nochunk (struct om
if (!gimple_omp_return_nowait_p (gsi_stmt (gsi)))
{
t = gimple_omp_return_lhs (gsi_stmt (gsi));
- gsi_insert_after (&gsi, omp_build_barrier (t), GSI_SAME_STMT);
+ if (fd->have_reductemp)
+ {
+ tree fn;
+ if (t)
+ fn = builtin_decl_explicit (BUILT_IN_GOMP_LOOP_END_CANCEL);
+ else
+ fn = builtin_decl_explicit (BUILT_IN_GOMP_LOOP_END);
+ gcall *g = gimple_build_call (fn, 0);
+ if (t)
+ {
+ gimple_call_set_lhs (g, t);
+ gsi_insert_after (&gsi, gimple_build_assign (reductions,
+ NOP_EXPR, t),
+ GSI_SAME_STMT);
+ }
+ gsi_insert_after (&gsi, g, GSI_SAME_STMT);
+ }
+ else
+ gsi_insert_after (&gsi, omp_build_barrier (t), GSI_SAME_STMT);
}
gsi_remove (&gsi, true);
@@ -3765,6 +3986,7 @@ expand_omp_for_static_chunk (struct omp_
bool broken_loop = region->cont == NULL;
tree *counts = NULL;
tree n1, n2, step;
+ tree reductions = NULL_TREE;
itype = type = TREE_TYPE (fd->loop.v);
if (POINTER_TYPE_P (type))
@@ -3852,6 +4074,29 @@ expand_omp_for_static_chunk (struct omp_
gsi = gsi_last_bb (entry_bb);
}
+ if (fd->have_reductemp)
+ {
+ tree t1 = build_int_cst (long_integer_type_node, 0);
+ tree t2 = build_int_cst (long_integer_type_node, 1);
+ tree t3 = build_int_cstu (long_integer_type_node,
+ (HOST_WIDE_INT_1U << 31) + 1);
+ tree clauses = gimple_omp_for_clauses (fd->for_stmt);
+ clauses = omp_find_clause (clauses, OMP_CLAUSE__REDUCTEMP_);
+ reductions = OMP_CLAUSE_DECL (clauses);
+ gcc_assert (TREE_CODE (reductions) == SSA_NAME);
+ gimple *g = SSA_NAME_DEF_STMT (reductions);
+ reductions = gimple_assign_rhs1 (g);
+ OMP_CLAUSE_DECL (clauses) = reductions;
+ gimple_stmt_iterator gsi2 = gsi_for_stmt (g);
+ tree t
+ = build_call_expr (builtin_decl_explicit (BUILT_IN_GOMP_LOOP_START),
+ 9, t1, t2, t2, t3, t1, null_pointer_node,
+ null_pointer_node, reductions, null_pointer_node);
+ force_gimple_operand_gsi (&gsi2, t, true, NULL_TREE,
+ true, GSI_SAME_STMT);
+ gsi_remove (&gsi2, true);
+ release_ssa_name (gimple_assign_lhs (g));
+ }
switch (gimple_omp_for_kind (fd->for_stmt))
{
case GF_OMP_FOR_KIND_FOR:
@@ -4155,7 +4400,25 @@ expand_omp_for_static_chunk (struct omp_
if (!gimple_omp_return_nowait_p (gsi_stmt (gsi)))
{
t = gimple_omp_return_lhs (gsi_stmt (gsi));
- gsi_insert_after (&gsi, omp_build_barrier (t), GSI_SAME_STMT);
+ if (fd->have_reductemp)
+ {
+ tree fn;
+ if (t)
+ fn = builtin_decl_explicit (BUILT_IN_GOMP_LOOP_END_CANCEL);
+ else
+ fn = builtin_decl_explicit (BUILT_IN_GOMP_LOOP_END);
+ gcall *g = gimple_build_call (fn, 0);
+ if (t)
+ {
+ gimple_call_set_lhs (g, t);
+ gsi_insert_after (&gsi, gimple_build_assign (reductions,
+ NOP_EXPR, t),
+ GSI_SAME_STMT);
+ }
+ gsi_insert_after (&gsi, g, GSI_SAME_STMT);
+ }
+ else
+ gsi_insert_after (&gsi, omp_build_barrier (t), GSI_SAME_STMT);
}
gsi_remove (&gsi, true);
@@ -5690,39 +5953,72 @@ expand_omp_for (struct omp_region *regio
else
{
int fn_index, start_ix, next_ix;
+ unsigned HOST_WIDE_INT sched = 0;
+ tree sched_arg = NULL_TREE;
gcc_assert (gimple_omp_for_kind (fd.for_stmt)
== GF_OMP_FOR_KIND_FOR);
if (fd.chunk_size == NULL
&& fd.sched_kind == OMP_CLAUSE_SCHEDULE_STATIC)
fd.chunk_size = integer_zero_node;
- gcc_assert (fd.sched_kind != OMP_CLAUSE_SCHEDULE_AUTO);
switch (fd.sched_kind)
{
case OMP_CLAUSE_SCHEDULE_RUNTIME:
- fn_index = 3;
+ if ((fd.sched_modifiers & OMP_CLAUSE_SCHEDULE_NONMONOTONIC) != 0)
+ {
+ gcc_assert (!fd.have_ordered);
+ fn_index = 6;
+ sched = 4;
+ }
+ else if ((fd.sched_modifiers & OMP_CLAUSE_SCHEDULE_MONOTONIC) == 0
+ && !fd.have_ordered)
+ fn_index = 7;
+ else
+ {
+ fn_index = 3;
+ sched = (HOST_WIDE_INT_1U << 31);
+ }
break;
case OMP_CLAUSE_SCHEDULE_DYNAMIC:
case OMP_CLAUSE_SCHEDULE_GUIDED:
- if ((fd.sched_modifiers & OMP_CLAUSE_SCHEDULE_NONMONOTONIC)
- && !fd.ordered
+ if ((fd.sched_modifiers & OMP_CLAUSE_SCHEDULE_MONOTONIC) == 0
&& !fd.have_ordered)
{
fn_index = 3 + fd.sched_kind;
+ sched = (fd.sched_kind == OMP_CLAUSE_SCHEDULE_GUIDED) + 2;
break;
}
- /* FALLTHRU */
- default:
fn_index = fd.sched_kind;
+ sched = (fd.sched_kind == OMP_CLAUSE_SCHEDULE_GUIDED) + 2;
+ sched += (HOST_WIDE_INT_1U << 31);
+ break;
+ case OMP_CLAUSE_SCHEDULE_STATIC:
+ gcc_assert (fd.have_ordered);
+ fn_index = 0;
+ sched = (HOST_WIDE_INT_1U << 31) + 1;
break;
+ default:
+ gcc_unreachable ();
}
if (!fd.ordered)
- fn_index += fd.have_ordered * 6;
+ fn_index += fd.have_ordered * 8;
if (fd.ordered)
start_ix = ((int)BUILT_IN_GOMP_LOOP_DOACROSS_STATIC_START) + fn_index;
else
start_ix = ((int)BUILT_IN_GOMP_LOOP_STATIC_START) + fn_index;
next_ix = ((int)BUILT_IN_GOMP_LOOP_STATIC_NEXT) + fn_index;
+ if (fd.have_reductemp)
+ {
+ if (fd.ordered)
+ start_ix = (int)BUILT_IN_GOMP_LOOP_DOACROSS_START;
+ else if (fd.have_ordered)
+ start_ix = (int)BUILT_IN_GOMP_LOOP_ORDERED_START;
+ else
+ start_ix = (int)BUILT_IN_GOMP_LOOP_START;
+ sched_arg = build_int_cstu (long_integer_type_node, sched);
+ if (!fd.chunk_size)
+ fd.chunk_size = integer_zero_node;
+ }
if (fd.iter_type == long_long_unsigned_type_node)
{
start_ix += ((int)BUILT_IN_GOMP_LOOP_ULL_STATIC_START
@@ -5731,7 +6027,8 @@ expand_omp_for (struct omp_region *regio
- (int)BUILT_IN_GOMP_LOOP_STATIC_NEXT);
}
expand_omp_for_generic (region, &fd, (enum built_in_function) start_ix,
- (enum built_in_function) next_ix, inner_stmt);
+ (enum built_in_function) next_ix, sched_arg,
+ inner_stmt);
}
if (gimple_in_ssa_p (cfun))
@@ -5831,7 +6128,25 @@ expand_omp_sections (struct omp_region *
sections_stmt = as_a <gomp_sections *> (gsi_stmt (si));
gcc_assert (gimple_code (sections_stmt) == GIMPLE_OMP_SECTIONS);
vin = gimple_omp_sections_control (sections_stmt);
- if (!is_combined_parallel (region))
+ tree clauses = gimple_omp_sections_clauses (sections_stmt);
+ tree reductmp = omp_find_clause (clauses, OMP_CLAUSE__REDUCTEMP_);
+ if (reductmp)
+ {
+ tree reductions = OMP_CLAUSE_DECL (reductmp);
+ gcc_assert (TREE_CODE (reductions) == SSA_NAME);
+ gimple *g = SSA_NAME_DEF_STMT (reductions);
+ reductions = gimple_assign_rhs1 (g);
+ OMP_CLAUSE_DECL (reductmp) = reductions;
+ gimple_stmt_iterator gsi = gsi_for_stmt (g);
+ t = build_int_cst (unsigned_type_node, len - 1);
+ u = builtin_decl_explicit (BUILT_IN_GOMP_SECTIONS2_START);
+ stmt = gimple_build_call (u, 3, t, reductions, null_pointer_node);
+ gimple_call_set_lhs (stmt, vin);
+ gsi_insert_before (&gsi, stmt, GSI_SAME_STMT);
+ gsi_remove (&gsi, true);
+ release_ssa_name (gimple_assign_lhs (g));
+ }
+ else if (!is_combined_parallel (region))
{
/* If we are not inside a combined parallel+sections region,
call GOMP_sections_start. */
@@ -5845,8 +6160,11 @@ expand_omp_sections (struct omp_region *
u = builtin_decl_explicit (BUILT_IN_GOMP_SECTIONS_NEXT);
stmt = gimple_build_call (u, 0);
}
- gimple_call_set_lhs (stmt, vin);
- gsi_insert_after (&si, stmt, GSI_SAME_STMT);
+ if (!reductmp)
+ {
+ gimple_call_set_lhs (stmt, vin);
+ gsi_insert_after (&si, stmt, GSI_SAME_STMT);
+ }
gsi_remove (&si, true);
/* The switch() statement replacing GIMPLE_OMP_SECTIONS_SWITCH goes in
@@ -6004,6 +6322,12 @@ expand_omp_synch (struct omp_region *reg
|| gimple_code (gsi_stmt (si)) == GIMPLE_OMP_ORDERED
|| gimple_code (gsi_stmt (si)) == GIMPLE_OMP_CRITICAL
|| gimple_code (gsi_stmt (si)) == GIMPLE_OMP_TEAMS);
+ if (gimple_code (gsi_stmt (si)) == GIMPLE_OMP_TEAMS
+ && gimple_omp_teams_host (as_a <gomp_teams *> (gsi_stmt (si))))
+ {
+ expand_omp_taskreg (region);
+ return;
+ }
gsi_remove (&si, true);
single_succ_edge (entry_bb)->flags = EDGE_FALLTHRU;
@@ -6016,6 +6340,24 @@ expand_omp_synch (struct omp_region *reg
}
}
+/* Translate enum omp_memory_order to enum memmodel. The two enums
+ are using different numbers so that OMP_MEMORY_ORDER_UNSPECIFIED
+ is 0. */
+
+static enum memmodel
+omp_memory_order_to_memmodel (enum omp_memory_order mo)
+{
+ switch (mo)
+ {
+ case OMP_MEMORY_ORDER_RELAXED: return MEMMODEL_RELAXED;
+ case OMP_MEMORY_ORDER_ACQUIRE: return MEMMODEL_ACQUIRE;
+ case OMP_MEMORY_ORDER_RELEASE: return MEMMODEL_RELEASE;
+ case OMP_MEMORY_ORDER_ACQ_REL: return MEMMODEL_ACQ_REL;
+ case OMP_MEMORY_ORDER_SEQ_CST: return MEMMODEL_SEQ_CST;
+ default: gcc_unreachable ();
+ }
+}
+
/* A subroutine of expand_omp_atomic. Attempt to implement the atomic
operation as a normal volatile load. */
@@ -6047,11 +6389,9 @@ expand_omp_atomic_load (basic_block load
type = TREE_TYPE (loaded_val);
itype = TREE_TYPE (TREE_TYPE (decl));
- call = build_call_expr_loc (loc, decl, 2, addr,
- build_int_cst (NULL,
- gimple_omp_atomic_seq_cst_p (stmt)
- ? MEMMODEL_SEQ_CST
- : MEMMODEL_RELAXED));
+ enum omp_memory_order omo = gimple_omp_atomic_memory_order (stmt);
+ tree mo = build_int_cst (NULL, omp_memory_order_to_memmodel (omo));
+ call = build_call_expr_loc (loc, decl, 2, addr, mo);
if (!useless_type_conversion_p (type, itype))
call = fold_build1_loc (loc, VIEW_CONVERT_EXPR, type, call);
call = build2_loc (loc, MODIFY_EXPR, void_type_node, loaded_val, call);
@@ -6122,11 +6462,9 @@ expand_omp_atomic_store (basic_block loa
if (!useless_type_conversion_p (itype, type))
stored_val = fold_build1_loc (loc, VIEW_CONVERT_EXPR, itype, stored_val);
- call = build_call_expr_loc (loc, decl, 3, addr, stored_val,
- build_int_cst (NULL,
- gimple_omp_atomic_seq_cst_p (stmt)
- ? MEMMODEL_SEQ_CST
- : MEMMODEL_RELAXED));
+ enum omp_memory_order omo = gimple_omp_atomic_memory_order (stmt);
+ tree mo = build_int_cst (NULL, omp_memory_order_to_memmodel (omo));
+ call = build_call_expr_loc (loc, decl, 3, addr, stored_val, mo);
if (exchange)
{
if (!useless_type_conversion_p (type, itype))
@@ -6167,7 +6505,6 @@ expand_omp_atomic_fetch_op (basic_block
enum tree_code code;
bool need_old, need_new;
machine_mode imode;
- bool seq_cst;
/* We expect to find the following sequences:
@@ -6200,7 +6537,9 @@ expand_omp_atomic_fetch_op (basic_block
return false;
need_new = gimple_omp_atomic_need_value_p (gsi_stmt (gsi));
need_old = gimple_omp_atomic_need_value_p (last_stmt (load_bb));
- seq_cst = gimple_omp_atomic_seq_cst_p (last_stmt (load_bb));
+ enum omp_memory_order omo
+ = gimple_omp_atomic_memory_order (last_stmt (load_bb));
+ enum memmodel mo = omp_memory_order_to_memmodel (omo);
gcc_checking_assert (!need_old || !need_new);
if (!operand_equal_p (gimple_assign_lhs (stmt), stored_val, 0))
@@ -6267,9 +6606,7 @@ expand_omp_atomic_fetch_op (basic_block
use the RELAXED memory model. */
call = build_call_expr_loc (loc, decl, 3, addr,
fold_convert_loc (loc, itype, rhs),
- build_int_cst (NULL,
- seq_cst ? MEMMODEL_SEQ_CST
- : MEMMODEL_RELAXED));
+ build_int_cst (NULL, mo));
if (need_old || need_new)
{
@@ -7921,6 +8258,10 @@ build_omp_regions_1 (basic_block bb, str
/* #pragma omp ordered depend is also just a stand-alone
directive. */
region = NULL;
+ else if (code == GIMPLE_OMP_TASK
+ && gimple_omp_task_taskwait_p (stmt))
+ /* #pragma omp taskwait depend(...) is a stand-alone directive. */
+ region = NULL;
/* ..., this directive becomes the parent for a new region. */
if (region)
parent = region;
@@ -8111,7 +8452,6 @@ omp_make_gimple_edges (basic_block bb, s
switch (code)
{
case GIMPLE_OMP_PARALLEL:
- case GIMPLE_OMP_TASK:
case GIMPLE_OMP_FOR:
case GIMPLE_OMP_SINGLE:
case GIMPLE_OMP_TEAMS:
@@ -8124,6 +8464,13 @@ omp_make_gimple_edges (basic_block bb, s
fallthru = true;
break;
+ case GIMPLE_OMP_TASK:
+ cur_region = new_omp_region (bb, code, cur_region);
+ fallthru = true;
+ if (gimple_omp_task_taskwait_p (last))
+ cur_region = cur_region->outer;
+ break;
+
case GIMPLE_OMP_ORDERED:
cur_region = new_omp_region (bb, code, cur_region);
fallthru = true;
@@ -36,6 +36,8 @@ along with GCC; see the file COPYING3.
#include "stringpool.h"
#include "attribs.h"
+enum omp_requires omp_requires_mask;
+
tree
omp_find_clause (tree clauses, enum omp_clause_code kind)
{
@@ -136,6 +138,7 @@ omp_extract_for_data (gomp_for *for_stmt
fd->pre = NULL;
fd->have_nowait = distribute || simd;
fd->have_ordered = false;
+ fd->have_reductemp = false;
fd->tiling = NULL_TREE;
fd->collapse = 1;
fd->ordered = 0;
@@ -186,6 +189,8 @@ omp_extract_for_data (gomp_for *for_stmt
collapse_iter = &OMP_CLAUSE_TILE_ITERVAR (t);
collapse_count = &OMP_CLAUSE_TILE_COUNT (t);
break;
+ case OMP_CLAUSE__REDUCTEMP_:
+ fd->have_reductemp = true;
default:
break;
}
@@ -250,13 +255,45 @@ omp_extract_for_data (gomp_for *for_stmt
loop->cond_code = gimple_omp_for_cond (for_stmt, i);
loop->n2 = gimple_omp_for_final (for_stmt, i);
- gcc_assert (loop->cond_code != NE_EXPR);
+ gcc_assert (loop->cond_code != NE_EXPR
+ || (gimple_omp_for_kind (for_stmt)
+ != GF_OMP_FOR_KIND_OACC_LOOP));
omp_adjust_for_condition (loc, &loop->cond_code, &loop->n2);
t = gimple_omp_for_incr (for_stmt, i);
gcc_assert (TREE_OPERAND (t, 0) == var);
loop->step = omp_get_for_step_from_incr (loc, t);
+ if (loop->cond_code == NE_EXPR)
+ {
+ gcc_assert (TREE_CODE (loop->step) == INTEGER_CST);
+ if (TREE_CODE (TREE_TYPE (loop->v)) == INTEGER_TYPE)
+ {
+ if (integer_onep (loop->step))
+ loop->cond_code = LT_EXPR;
+ else
+ {
+ gcc_assert (integer_minus_onep (loop->step));
+ loop->cond_code = GT_EXPR;
+ }
+ }
+ else
+ {
+ tree unit = TYPE_SIZE_UNIT (TREE_TYPE (TREE_TYPE (loop->v)));
+ gcc_assert (TREE_CODE (unit) == INTEGER_CST);
+ if (tree_int_cst_equal (unit, loop->step))
+ loop->cond_code = LT_EXPR;
+ else
+ {
+ gcc_assert (wi::neg (wi::to_widest (unit))
+ == wi::to_widest (loop->step));
+ loop->cond_code = GT_EXPR;
+ }
+ }
+ }
+
+ omp_adjust_for_condition (loc, &loop->cond_code, &loop->n2);
+
if (simd
|| (fd->sched_kind == OMP_CLAUSE_SCHEDULE_STATIC
&& !fd->have_ordered))
@@ -281,9 +318,8 @@ omp_extract_for_data (gomp_for *for_stmt
tree n;
if (loop->cond_code == LT_EXPR)
- n = fold_build2_loc (loc,
- PLUS_EXPR, TREE_TYPE (loop->v),
- loop->n2, loop->step);
+ n = fold_build2_loc (loc, PLUS_EXPR, TREE_TYPE (loop->v),
+ loop->n2, loop->step);
else
n = loop->n1;
if (TREE_CODE (n) != INTEGER_CST
@@ -298,15 +334,13 @@ omp_extract_for_data (gomp_for *for_stmt
if (loop->cond_code == LT_EXPR)
{
n1 = loop->n1;
- n2 = fold_build2_loc (loc,
- PLUS_EXPR, TREE_TYPE (loop->v),
- loop->n2, loop->step);
+ n2 = fold_build2_loc (loc, PLUS_EXPR, TREE_TYPE (loop->v),
+ loop->n2, loop->step);
}
else
{
- n1 = fold_build2_loc (loc,
- MINUS_EXPR, TREE_TYPE (loop->v),
- loop->n2, loop->step);
+ n1 = fold_build2_loc (loc, MINUS_EXPR, TREE_TYPE (loop->v),
+ loop->n2, loop->step);
n2 = loop->n1;
}
if (TREE_CODE (n1) != INTEGER_CST
@@ -338,27 +372,31 @@ omp_extract_for_data (gomp_for *for_stmt
if (POINTER_TYPE_P (itype))
itype = signed_type_for (itype);
t = build_int_cst (itype, (loop->cond_code == LT_EXPR ? -1 : 1));
- t = fold_build2_loc (loc,
- PLUS_EXPR, itype,
- fold_convert_loc (loc, itype, loop->step), t);
+ t = fold_build2_loc (loc, PLUS_EXPR, itype,
+ fold_convert_loc (loc, itype, loop->step),
+ t);
t = fold_build2_loc (loc, PLUS_EXPR, itype, t,
- fold_convert_loc (loc, itype, loop->n2));
+ fold_convert_loc (loc, itype, loop->n2));
t = fold_build2_loc (loc, MINUS_EXPR, itype, t,
- fold_convert_loc (loc, itype, loop->n1));
+ fold_convert_loc (loc, itype, loop->n1));
if (TYPE_UNSIGNED (itype) && loop->cond_code == GT_EXPR)
- t = fold_build2_loc (loc, TRUNC_DIV_EXPR, itype,
- fold_build1_loc (loc, NEGATE_EXPR, itype, t),
- fold_build1_loc (loc, NEGATE_EXPR, itype,
- fold_convert_loc (loc, itype,
- loop->step)));
+ {
+ tree step = fold_convert_loc (loc, itype, loop->step);
+ t = fold_build2_loc (loc, TRUNC_DIV_EXPR, itype,
+ fold_build1_loc (loc, NEGATE_EXPR,
+ itype, t),
+ fold_build1_loc (loc, NEGATE_EXPR,
+ itype, step));
+ }
else
t = fold_build2_loc (loc, TRUNC_DIV_EXPR, itype, t,
- fold_convert_loc (loc, itype, loop->step));
+ fold_convert_loc (loc, itype,
+ loop->step));
t = fold_convert_loc (loc, long_long_unsigned_type_node, t);
if (count != NULL_TREE)
- count = fold_build2_loc (loc,
- MULT_EXPR, long_long_unsigned_type_node,
- count, t);
+ count = fold_build2_loc (loc, MULT_EXPR,
+ long_long_unsigned_type_node,
+ count, t);
else
count = t;
if (TREE_CODE (count) != INTEGER_CST)
@@ -62,7 +62,7 @@ struct omp_for_data
tree tiling; /* Tiling values (if non null). */
int collapse; /* Collapsed loops, 1 for a non-collapsed loop. */
int ordered;
- bool have_nowait, have_ordered, simd_schedule;
+ bool have_nowait, have_ordered, simd_schedule, have_reductemp;
unsigned char sched_modifiers;
enum omp_clause_schedule_kind sched_kind;
struct omp_for_data_loop *loops;
@@ -89,4 +89,16 @@ extern bool offloading_function_p (tree
extern int oacc_get_fn_dim_size (tree fn, int axis);
extern int oacc_get_ifn_dim_arg (const gimple *stmt);
+enum omp_requires {
+ OMP_REQUIRES_ATOMIC_DEFAULT_MEM_ORDER = 0xf,
+ OMP_REQUIRES_UNIFIED_ADDRESS = 0x10,
+ OMP_REQUIRES_UNIFIED_SHARED_MEMORY = 0x20,
+ OMP_REQUIRES_DYNAMIC_ALLOCATORS = 0x40,
+ OMP_REQUIRES_REVERSE_OFFLOAD = 0x80,
+ OMP_REQUIRES_ATOMIC_DEFAULT_MEM_ORDER_USED = 0x100,
+ OMP_REQUIRES_TARGET_USED = 0x200
+};
+
+extern GTY(()) enum omp_requires omp_requires_mask;
+
#endif /* GCC_OMP_GENERAL_H */
@@ -1053,8 +1053,8 @@ grid_eliminate_combined_simd_part (gomp_
while (*tgt)
tgt = &OMP_CLAUSE_CHAIN (*tgt);
- /* Copy over all clauses, except for linaer clauses, which are turned into
- private clauses, and all other simd-specificl clauses, which are
+ /* Copy over all clauses, except for linear clauses, which are turned into
+ private clauses, and all other simd-specific clauses, which are
ignored. */
tree *pc = gimple_omp_for_clauses_ptr (simd);
while (*pc)
@@ -1083,7 +1083,7 @@ grid_eliminate_combined_simd_part (gomp_
*pc = OMP_CLAUSE_CHAIN (c);
OMP_CLAUSE_CHAIN (c) = NULL;
*tgt = c;
- tgt = &OMP_CLAUSE_CHAIN(c);
+ tgt = &OMP_CLAUSE_CHAIN (c);
break;
}
}
@@ -114,6 +114,15 @@ struct omp_context
otherwise. */
gimple *simt_stmt;
+ /* For task reductions registered in this context, a vector containing
+ the length of the private copies block (if constant, otherwise NULL)
+ and then offsets (if constant, otherwise NULL) for each entry. */
+ vec<tree> task_reductions;
+
+ /* And a hash map from the reduction clauses to the registered array
+ elts. */
+ hash_map<tree, unsigned> *task_reduction_map;
+
/* Nesting depth of this context. Used to beautify error messages re
invalid gotos. The outermost ctx is depth 1, with depth 0 being
reserved for the main body of the function. */
@@ -280,12 +289,23 @@ is_taskloop_ctx (omp_context *ctx)
}
-/* Return true if CTX is for an omp parallel or omp task. */
+/* Return true if CTX is for a host omp teams. */
+
+static inline bool
+is_host_teams_ctx (omp_context *ctx)
+{
+ return gimple_code (ctx->stmt) == GIMPLE_OMP_TEAMS
+ && gimple_omp_teams_host (as_a <gomp_teams *> (ctx->stmt));
+}
+
+/* Return true if CTX is for an omp parallel or omp task or host omp teams
+ (the last one is strictly not a task region in OpenMP speak, but we
+ need to treat it similarly). */
static inline bool
is_taskreg_ctx (omp_context *ctx)
{
- return is_parallel_ctx (ctx) || is_task_ctx (ctx);
+ return is_parallel_ctx (ctx) || is_task_ctx (ctx) || is_host_teams_ctx (ctx);
}
/* Return true if EXPR is variable sized. */
@@ -371,7 +391,7 @@ use_pointer_for_field (tree decl, omp_co
be passing an address in this case? Should we simply assert
this to be false, or should we have a cleanup pass that removes
these from the list of mappings? */
- if (TREE_STATIC (decl) || DECL_EXTERNAL (decl))
+ if (is_global_var (maybe_lookup_decl_in_outer_ctx (decl, shared_ctx)))
return true;
/* For variables with DECL_HAS_VALUE_EXPR_P set, we cannot tell
@@ -522,6 +542,9 @@ build_outer_var_ref (tree var, omp_conte
enum omp_clause_code code = OMP_CLAUSE_ERROR)
{
tree x;
+ omp_context *outer = ctx->outer;
+ while (outer && gimple_code (outer->stmt) == GIMPLE_OMP_TASKGROUP)
+ outer = outer->outer;
if (is_global_var (maybe_lookup_decl_in_outer_ctx (var, ctx)))
x = var;
@@ -548,44 +571,43 @@ build_outer_var_ref (tree var, omp_conte
Similarly for OMP_CLAUSE_PRIVATE with outer ref, that can refer
to private vars in all worksharing constructs. */
x = NULL_TREE;
- if (ctx->outer && is_taskreg_ctx (ctx))
- x = lookup_decl (var, ctx->outer);
- else if (ctx->outer)
+ if (outer && is_taskreg_ctx (outer))
+ x = lookup_decl (var, outer);
+ else if (outer)
x = maybe_lookup_decl_in_outer_ctx (var, ctx);
if (x == NULL_TREE)
x = var;
}
else if (code == OMP_CLAUSE_LASTPRIVATE && is_taskloop_ctx (ctx))
{
- gcc_assert (ctx->outer);
+ gcc_assert (outer);
splay_tree_node n
- = splay_tree_lookup (ctx->outer->field_map,
+ = splay_tree_lookup (outer->field_map,
(splay_tree_key) &DECL_UID (var));
if (n == NULL)
{
- if (is_global_var (maybe_lookup_decl_in_outer_ctx (var, ctx->outer)))
+ if (is_global_var (maybe_lookup_decl_in_outer_ctx (var, outer)))
x = var;
else
- x = lookup_decl (var, ctx->outer);
+ x = lookup_decl (var, outer);
}
else
{
tree field = (tree) n->value;
/* If the receiver record type was remapped in the child function,
remap the field into the new record type. */
- x = maybe_lookup_field (field, ctx->outer);
+ x = maybe_lookup_field (field, outer);
if (x != NULL)
field = x;
- x = build_simple_mem_ref (ctx->outer->receiver_decl);
+ x = build_simple_mem_ref (outer->receiver_decl);
x = omp_build_component_ref (x, field);
- if (use_pointer_for_field (var, ctx->outer))
+ if (use_pointer_for_field (var, outer))
x = build_simple_mem_ref (x);
}
}
- else if (ctx->outer)
+ else if (outer)
{
- omp_context *outer = ctx->outer;
if (gimple_code (outer->stmt) == GIMPLE_OMP_GRID_BODY)
{
outer = outer->outer;
@@ -925,6 +947,12 @@ delete_omp_context (splay_tree_value val
if (is_task_ctx (ctx))
finalize_task_copyfn (as_a <gomp_task *> (ctx->stmt));
+ if (ctx->task_reduction_map)
+ {
+ ctx->task_reductions.release ();
+ delete ctx->task_reduction_map;
+ }
+
XDELETE (ctx);
}
@@ -1011,8 +1039,10 @@ scan_sharing_clauses (tree clauses, omp_
case OMP_CLAUSE_SHARED:
decl = OMP_CLAUSE_DECL (c);
- /* Ignore shared directives in teams construct. */
- if (gimple_code (ctx->stmt) == GIMPLE_OMP_TEAMS)
+ /* Ignore shared directives in teams construct inside of
+ target construct. */
+ if (gimple_code (ctx->stmt) == GIMPLE_OMP_TEAMS
+ && !is_host_teams_ctx (ctx))
{
/* Global variables don't need to be copied,
the receiver side will use them directly. */
@@ -1050,9 +1080,9 @@ scan_sharing_clauses (tree clauses, omp_
goto do_private;
case OMP_CLAUSE_REDUCTION:
+ case OMP_CLAUSE_IN_REDUCTION:
decl = OMP_CLAUSE_DECL (c);
- if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_REDUCTION
- && TREE_CODE (decl) == MEM_REF)
+ if (TREE_CODE (decl) == MEM_REF)
{
tree t = TREE_OPERAND (decl, 0);
if (TREE_CODE (t) == POINTER_PLUS_EXPR)
@@ -1062,12 +1092,50 @@ scan_sharing_clauses (tree clauses, omp_
t = TREE_OPERAND (t, 0);
install_var_local (t, ctx);
if (is_taskreg_ctx (ctx)
- && !is_global_var (maybe_lookup_decl_in_outer_ctx (t, ctx))
- && !is_variable_sized (t))
+ && (!is_global_var (maybe_lookup_decl_in_outer_ctx (t, ctx))
+ || (is_task_ctx (ctx)
+ && (TREE_CODE (TREE_TYPE (t)) == POINTER_TYPE
+ || (TREE_CODE (TREE_TYPE (t)) == REFERENCE_TYPE
+ && (TREE_CODE (TREE_TYPE (TREE_TYPE (t)))
+ == POINTER_TYPE)))))
+ && !is_variable_sized (t)
+ && (OMP_CLAUSE_CODE (c) != OMP_CLAUSE_REDUCTION
+ || (!OMP_CLAUSE_REDUCTION_TASK (c)
+ && !is_task_ctx (ctx))))
+ {
+ by_ref = use_pointer_for_field (t, NULL);
+ if (is_task_ctx (ctx)
+ && TREE_CODE (TREE_TYPE (t)) == REFERENCE_TYPE
+ && TREE_CODE (TREE_TYPE (TREE_TYPE (t))) == POINTER_TYPE)
+ {
+ install_var_field (t, false, 1, ctx);
+ install_var_field (t, by_ref, 2, ctx);
+ }
+ else
+ install_var_field (t, by_ref, 3, ctx);
+ }
+ break;
+ }
+ if (is_task_ctx (ctx)
+ || (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_REDUCTION
+ && OMP_CLAUSE_REDUCTION_TASK (c)
+ && is_parallel_ctx (ctx)))
+ {
+ /* Global variables don't need to be copied,
+ the receiver side will use them directly. */
+ if (!is_global_var (maybe_lookup_decl_in_outer_ctx (decl, ctx)))
{
- by_ref = use_pointer_for_field (t, ctx);
- install_var_field (t, by_ref, 3, ctx);
+ by_ref = use_pointer_for_field (decl, ctx);
+ if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_IN_REDUCTION)
+ install_var_field (decl, by_ref, 3, ctx);
}
+ install_var_local (decl, ctx);
+ break;
+ }
+ if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_REDUCTION
+ && OMP_CLAUSE_REDUCTION_TASK (c))
+ {
+ install_var_local (decl, ctx);
break;
}
goto do_private;
@@ -1142,6 +1210,7 @@ scan_sharing_clauses (tree clauses, omp_
goto do_private;
case OMP_CLAUSE__LOOPTEMP_:
+ case OMP_CLAUSE__REDUCTEMP_:
gcc_assert (is_taskreg_ctx (ctx));
decl = OMP_CLAUSE_DECL (c);
install_var_field (decl, false, 3, ctx);
@@ -1323,8 +1392,10 @@ scan_sharing_clauses (tree clauses, omp_
case OMP_CLAUSE_TILE:
case OMP_CLAUSE__SIMT_:
case OMP_CLAUSE_DEFAULT:
+ case OMP_CLAUSE_NONTEMPORAL:
case OMP_CLAUSE_IF_PRESENT:
case OMP_CLAUSE_FINALIZE:
+ case OMP_CLAUSE_TASK_REDUCTION:
break;
case OMP_CLAUSE_ALIGNED:
@@ -1382,6 +1453,7 @@ scan_sharing_clauses (tree clauses, omp_
break;
case OMP_CLAUSE_REDUCTION:
+ case OMP_CLAUSE_IN_REDUCTION:
decl = OMP_CLAUSE_DECL (c);
if (TREE_CODE (decl) != MEM_REF)
{
@@ -1393,9 +1465,16 @@ scan_sharing_clauses (tree clauses, omp_
scan_array_reductions = true;
break;
+ case OMP_CLAUSE_TASK_REDUCTION:
+ if (OMP_CLAUSE_REDUCTION_PLACEHOLDER (c))
+ scan_array_reductions = true;
+ break;
+
case OMP_CLAUSE_SHARED:
- /* Ignore shared directives in teams construct. */
- if (gimple_code (ctx->stmt) == GIMPLE_OMP_TEAMS)
+ /* Ignore shared directives in teams construct inside of
+ target construct. */
+ if (gimple_code (ctx->stmt) == GIMPLE_OMP_TEAMS
+ && !is_host_teams_ctx (ctx))
break;
decl = OMP_CLAUSE_DECL (c);
if (is_global_var (maybe_lookup_decl_in_outer_ctx (decl, ctx)))
@@ -1472,6 +1551,7 @@ scan_sharing_clauses (tree clauses, omp_
case OMP_CLAUSE_ALIGNED:
case OMP_CLAUSE_DEPEND:
case OMP_CLAUSE__LOOPTEMP_:
+ case OMP_CLAUSE__REDUCTEMP_:
case OMP_CLAUSE_TO:
case OMP_CLAUSE_FROM:
case OMP_CLAUSE_PRIORITY:
@@ -1482,6 +1562,7 @@ scan_sharing_clauses (tree clauses, omp_
case OMP_CLAUSE_NOGROUP:
case OMP_CLAUSE_DEFAULTMAP:
case OMP_CLAUSE_USE_DEVICE_PTR:
+ case OMP_CLAUSE_NONTEMPORAL:
case OMP_CLAUSE_ASYNC:
case OMP_CLAUSE_WAIT:
case OMP_CLAUSE_NUM_GANGS:
@@ -1511,7 +1592,9 @@ scan_sharing_clauses (tree clauses, omp_
if (scan_array_reductions)
{
for (c = clauses; c; c = OMP_CLAUSE_CHAIN (c))
- if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_REDUCTION
+ if ((OMP_CLAUSE_CODE (c) == OMP_CLAUSE_REDUCTION
+ || OMP_CLAUSE_CODE (c) == OMP_CLAUSE_IN_REDUCTION
+ || OMP_CLAUSE_CODE (c) == OMP_CLAUSE_TASK_REDUCTION)
&& OMP_CLAUSE_REDUCTION_PLACEHOLDER (c))
{
scan_omp (&OMP_CLAUSE_REDUCTION_GIMPLE_INIT (c), ctx);
@@ -1700,7 +1783,7 @@ omp_find_combined_for (gimple_stmt_itera
return NULL;
}
-/* Add _LOOPTEMP_ clauses on OpenMP parallel or task. */
+/* Add _LOOPTEMP_/_REDUCTEMP_ clauses on OpenMP parallel or task. */
static void
add_taskreg_looptemp_clauses (enum gf_mask msk, gimple *stmt,
@@ -1747,6 +1830,18 @@ add_taskreg_looptemp_clauses (enum gf_ma
gimple_omp_taskreg_set_clauses (stmt, c);
}
}
+ if (msk == GF_OMP_FOR_KIND_TASKLOOP
+ && omp_find_clause (gimple_omp_task_clauses (stmt),
+ OMP_CLAUSE_REDUCTION))
+ {
+ tree type = build_pointer_type (pointer_sized_int_node);
+ tree temp = create_tmp_var (type);
+ tree c = build_omp_clause (UNKNOWN_LOCATION, OMP_CLAUSE__REDUCTEMP_);
+ insert_decl_map (&outer_ctx->cb, temp, temp);
+ OMP_CLAUSE_DECL (c) = temp;
+ OMP_CLAUSE_CHAIN (c) = gimple_omp_task_clauses (stmt);
+ gimple_omp_task_set_clauses (stmt, c);
+ }
}
/* Scan an OpenMP parallel directive. */
@@ -1771,6 +1866,23 @@ scan_omp_parallel (gimple_stmt_iterator
if (gimple_omp_parallel_combined_p (stmt))
add_taskreg_looptemp_clauses (GF_OMP_FOR_KIND_FOR, stmt, outer_ctx);
+ for (tree c = omp_find_clause (gimple_omp_parallel_clauses (stmt),
+ OMP_CLAUSE_REDUCTION);
+ c; c = omp_find_clause (OMP_CLAUSE_CHAIN (c), OMP_CLAUSE_REDUCTION))
+ if (OMP_CLAUSE_REDUCTION_TASK (c))
+ {
+ tree type = build_pointer_type (pointer_sized_int_node);
+ tree temp = create_tmp_var (type);
+ tree c = build_omp_clause (UNKNOWN_LOCATION, OMP_CLAUSE__REDUCTEMP_);
+ if (outer_ctx)
+ insert_decl_map (&outer_ctx->cb, temp, temp);
+ OMP_CLAUSE_DECL (c) = temp;
+ OMP_CLAUSE_CHAIN (c) = gimple_omp_parallel_clauses (stmt);
+ gimple_omp_parallel_set_clauses (stmt, c);
+ break;
+ }
+ else if (OMP_CLAUSE_CHAIN (c) == NULL_TREE)
+ break;
ctx = new_omp_context (stmt, outer_ctx);
taskreg_contexts.safe_push (ctx);
@@ -1810,6 +1922,7 @@ scan_omp_task (gimple_stmt_iterator *gsi
/* Ignore task directives with empty bodies, unless they have depend
clause. */
if (optimize > 0
+ && gimple_omp_body (stmt)
&& empty_body_p (gimple_omp_body (stmt))
&& !omp_find_clause (gimple_omp_task_clauses (stmt), OMP_CLAUSE_DEPEND))
{
@@ -1821,6 +1934,13 @@ scan_omp_task (gimple_stmt_iterator *gsi
add_taskreg_looptemp_clauses (GF_OMP_FOR_KIND_TASKLOOP, stmt, outer_ctx);
ctx = new_omp_context (stmt, outer_ctx);
+
+ if (gimple_omp_task_taskwait_p (stmt))
+ {
+ scan_sharing_clauses (gimple_omp_task_clauses (stmt), ctx);
+ return;
+ }
+
taskreg_contexts.safe_push (ctx);
if (taskreg_nesting_level > 1)
ctx->is_nested = true;
@@ -1897,7 +2017,7 @@ finish_taskreg_scan (omp_context *ctx)
return;
/* If any task_shared_vars were needed, verify all
- OMP_CLAUSE_SHARED clauses on GIMPLE_OMP_{PARALLEL,TASK}
+ OMP_CLAUSE_SHARED clauses on GIMPLE_OMP_{PARALLEL,TASK,TEAMS}
statements if use_pointer_for_field hasn't changed
because of that. If it did, update field types now. */
if (task_shared_vars)
@@ -1943,6 +2063,30 @@ finish_taskreg_scan (omp_context *ctx)
if (gimple_code (ctx->stmt) == GIMPLE_OMP_PARALLEL)
{
+ tree clauses = gimple_omp_parallel_clauses (ctx->stmt);
+ tree c = omp_find_clause (clauses, OMP_CLAUSE__REDUCTEMP_);
+ if (c)
+ {
+ /* Move the _reductemp_ clause first. GOMP_parallel_reductions
+ expects to find it at the start of data. */
+ tree f = lookup_field (OMP_CLAUSE_DECL (c), ctx);
+ tree *p = &TYPE_FIELDS (ctx->record_type);
+ while (*p)
+ if (*p == f)
+ {
+ *p = DECL_CHAIN (*p);
+ break;
+ }
+ else
+ p = &DECL_CHAIN (*p);
+ DECL_CHAIN (f) = TYPE_FIELDS (ctx->record_type);
+ TYPE_FIELDS (ctx->record_type) = f;
+ }
+ layout_type (ctx->record_type);
+ fixup_child_record_type (ctx);
+ }
+ else if (gimple_code (ctx->stmt) == GIMPLE_OMP_TEAMS)
+ {
layout_type (ctx->record_type);
fixup_child_record_type (ctx);
}
@@ -1969,33 +2113,50 @@ finish_taskreg_scan (omp_context *ctx)
/* Move fields corresponding to first and second _looptemp_
clause first. There are filled by GOMP_taskloop
and thus need to be in specific positions. */
- tree c1 = gimple_omp_task_clauses (ctx->stmt);
- c1 = omp_find_clause (c1, OMP_CLAUSE__LOOPTEMP_);
+ tree clauses = gimple_omp_task_clauses (ctx->stmt);
+ tree c1 = omp_find_clause (clauses, OMP_CLAUSE__LOOPTEMP_);
tree c2 = omp_find_clause (OMP_CLAUSE_CHAIN (c1),
OMP_CLAUSE__LOOPTEMP_);
+ tree c3 = omp_find_clause (clauses, OMP_CLAUSE__REDUCTEMP_);
tree f1 = lookup_field (OMP_CLAUSE_DECL (c1), ctx);
tree f2 = lookup_field (OMP_CLAUSE_DECL (c2), ctx);
+ tree f3 = c3 ? lookup_field (OMP_CLAUSE_DECL (c3), ctx) : NULL_TREE;
p = &TYPE_FIELDS (ctx->record_type);
while (*p)
- if (*p == f1 || *p == f2)
+ if (*p == f1 || *p == f2 || *p == f3)
*p = DECL_CHAIN (*p);
else
p = &DECL_CHAIN (*p);
DECL_CHAIN (f1) = f2;
- DECL_CHAIN (f2) = TYPE_FIELDS (ctx->record_type);
+ if (c3)
+ {
+ DECL_CHAIN (f2) = f3;
+ DECL_CHAIN (f3) = TYPE_FIELDS (ctx->record_type);
+ }
+ else
+ DECL_CHAIN (f2) = TYPE_FIELDS (ctx->record_type);
TYPE_FIELDS (ctx->record_type) = f1;
if (ctx->srecord_type)
{
f1 = lookup_sfield (OMP_CLAUSE_DECL (c1), ctx);
f2 = lookup_sfield (OMP_CLAUSE_DECL (c2), ctx);
+ if (c3)
+ f3 = lookup_sfield (OMP_CLAUSE_DECL (c3), ctx);
p = &TYPE_FIELDS (ctx->srecord_type);
while (*p)
- if (*p == f1 || *p == f2)
+ if (*p == f1 || *p == f2 || *p == f3)
*p = DECL_CHAIN (*p);
else
p = &DECL_CHAIN (*p);
DECL_CHAIN (f1) = f2;
DECL_CHAIN (f2) = TYPE_FIELDS (ctx->srecord_type);
+ if (c3)
+ {
+ DECL_CHAIN (f2) = f3;
+ DECL_CHAIN (f3) = TYPE_FIELDS (ctx->srecord_type);
+ }
+ else
+ DECL_CHAIN (f2) = TYPE_FIELDS (ctx->srecord_type);
TYPE_FIELDS (ctx->srecord_type) = f1;
}
}
@@ -2154,7 +2315,7 @@ scan_omp_for (gomp_for *stmt, omp_contex
if (tgt && is_oacc_kernels (tgt))
{
- /* Strip out reductions, as they are not handled yet. */
+ /* Strip out reductions, as they are not handled yet. */
tree *prev_ptr = &clauses;
while (tree probe = *prev_ptr)
@@ -2321,8 +2482,32 @@ static void
scan_omp_teams (gomp_teams *stmt, omp_context *outer_ctx)
{
omp_context *ctx = new_omp_context (stmt, outer_ctx);
+
+ if (!gimple_omp_teams_host (stmt))
+ {
+ scan_sharing_clauses (gimple_omp_teams_clauses (stmt), ctx);
+ scan_omp (gimple_omp_body_ptr (stmt), ctx);
+ return;
+ }
+ taskreg_contexts.safe_push (ctx);
+ gcc_assert (taskreg_nesting_level == 1);
+ ctx->field_map = splay_tree_new (splay_tree_compare_pointers, 0, 0);
+ ctx->record_type = lang_hooks.types.make_type (RECORD_TYPE);
+ tree name = create_tmp_var_name (".omp_data_s");
+ name = build_decl (gimple_location (stmt),
+ TYPE_DECL, name, ctx->record_type);
+ DECL_ARTIFICIAL (name) = 1;
+ DECL_NAMELESS (name) = 1;
+ TYPE_NAME (ctx->record_type) = name;
+ TYPE_ARTIFICIAL (ctx->record_type) = 1;
+ create_omp_child_function (ctx, false);
+ gimple_omp_teams_set_child_fn (stmt, ctx->cb.dst_fn);
+
scan_sharing_clauses (gimple_omp_teams_clauses (stmt), ctx);
scan_omp (gimple_omp_body_ptr (stmt), ctx);
+
+ if (TYPE_FIELDS (ctx->record_type) == NULL)
+ ctx->record_type = ctx->receiver_decl = NULL;
}
/* Check nesting restrictions. */
@@ -2388,9 +2573,13 @@ check_omp_nesting_restrictions (gimple *
return true;
}
}
+ else if (gimple_code (stmt) == GIMPLE_OMP_ATOMIC_LOAD
+ || gimple_code (stmt) == GIMPLE_OMP_ATOMIC_STORE)
+ return true;
error_at (gimple_location (stmt),
"OpenMP constructs other than %<#pragma omp ordered simd%>"
- " may not be nested inside %<simd%> region");
+ " or %<#pragma omp atomic%> may not be nested inside"
+ " %<simd%> region");
return false;
}
else if (gimple_code (ctx->stmt) == GIMPLE_OMP_TEAMS)
@@ -2814,13 +3003,20 @@ check_omp_nesting_restrictions (gimple *
}
break;
case GIMPLE_OMP_TEAMS:
- if (ctx == NULL
- || gimple_code (ctx->stmt) != GIMPLE_OMP_TARGET
- || gimple_omp_target_kind (ctx->stmt) != GF_OMP_TARGET_KIND_REGION)
- {
+ if (ctx == NULL)
+ break;
+ else if (gimple_code (ctx->stmt) != GIMPLE_OMP_TARGET
+ || (gimple_omp_target_kind (ctx->stmt)
+ != GF_OMP_TARGET_KIND_REGION))
+ {
+ /* Teams construct can appear either strictly nested inside of
+ target construct with no intervening stmts, or can be encountered
+ only by initial task (so must not appear inside any OpenMP
+ construct. */
error_at (gimple_location (stmt),
- "%<teams%> construct not closely nested inside of "
- "%<target%> construct");
+ "%<teams%> construct must be closely nested inside of "
+ "%<target%> construct or not nested in any OpenMP "
+ "construct");
return false;
}
break;
@@ -3090,7 +3286,6 @@ scan_omp_1_stmt (gimple_stmt_iterator *g
case GIMPLE_OMP_SECTION:
case GIMPLE_OMP_MASTER:
- case GIMPLE_OMP_TASKGROUP:
case GIMPLE_OMP_ORDERED:
case GIMPLE_OMP_CRITICAL:
case GIMPLE_OMP_GRID_BODY:
@@ -3098,12 +3293,25 @@ scan_omp_1_stmt (gimple_stmt_iterator *g
scan_omp (gimple_omp_body_ptr (stmt), ctx);
break;
+ case GIMPLE_OMP_TASKGROUP:
+ ctx = new_omp_context (stmt, ctx);
+ scan_sharing_clauses (gimple_omp_taskgroup_clauses (stmt), ctx);
+ scan_omp (gimple_omp_body_ptr (stmt), ctx);
+ break;
+
case GIMPLE_OMP_TARGET:
scan_omp_target (as_a <gomp_target *> (stmt), ctx);
break;
case GIMPLE_OMP_TEAMS:
- scan_omp_teams (as_a <gomp_teams *> (stmt), ctx);
+ if (gimple_omp_teams_host (as_a <gomp_teams *> (stmt)))
+ {
+ taskreg_nesting_level++;
+ scan_omp_teams (as_a <gomp_teams *> (stmt), ctx);
+ taskreg_nesting_level--;
+ }
+ else
+ scan_omp_teams (as_a <gomp_teams *> (stmt), ctx);
break;
case GIMPLE_BIND:
@@ -3515,6 +3723,30 @@ handle_simd_reference (location_t loc, t
}
}
+/* Helper function for lower_rec_input_clauses. Emit into ilist sequence
+ code to emit (type) (tskred_temp[idx]). */
+
+static tree
+task_reduction_read (gimple_seq *ilist, tree tskred_temp, tree type,
+ unsigned idx)
+{
+ unsigned HOST_WIDE_INT sz
+ = tree_to_uhwi (TYPE_SIZE_UNIT (pointer_sized_int_node));
+ tree r = build2 (MEM_REF, pointer_sized_int_node,
+ tskred_temp, build_int_cst (TREE_TYPE (tskred_temp),
+ idx * sz));
+ tree v = create_tmp_var (pointer_sized_int_node);
+ gimple *g = gimple_build_assign (v, r);
+ gimple_seq_add_stmt (ilist, g);
+ if (!useless_type_conversion_p (type, pointer_sized_int_node))
+ {
+ v = create_tmp_var (type);
+ g = gimple_build_assign (v, NOP_EXPR, gimple_assign_lhs (g));
+ gimple_seq_add_stmt (ilist, g);
+ }
+ return v;
+}
+
/* Generate code to implement the input clauses, FIRSTPRIVATE and COPYIN,
from the receiver (aka child) side and initializers for REFERENCE_TYPE
private variables. Initialization statements go in ILIST, while calls
@@ -3558,6 +3790,7 @@ lower_rec_input_clauses (tree clauses, g
sctx.max_vf = 1;
break;
case OMP_CLAUSE_REDUCTION:
+ case OMP_CLAUSE_IN_REDUCTION:
if (TREE_CODE (OMP_CLAUSE_DECL (c)) == MEM_REF
|| is_variable_sized (OMP_CLAUSE_DECL (c)))
sctx.max_vf = 1;
@@ -3570,18 +3803,87 @@ lower_rec_input_clauses (tree clauses, g
if (sctx.is_simt && maybe_ne (sctx.max_vf, 1U))
sctx.simt_eargs.safe_push (NULL_TREE);
+ unsigned task_reduction_cnt = 0;
+ unsigned task_reduction_cntorig = 0;
+ unsigned task_reduction_cnt_full = 0;
+ unsigned task_reduction_cntorig_full = 0;
+ unsigned task_reduction_other_cnt = 0;
+ tree tskred_atype = NULL_TREE, tskred_avar = NULL_TREE;
+ tree tskred_base = NULL_TREE, tskred_temp = NULL_TREE;
/* Do all the fixed sized types in the first pass, and the variable sized
types in the second pass. This makes sure that the scalar arguments to
the variable sized types are processed before we use them in the
- variable sized operations. */
- for (pass = 0; pass < 2; ++pass)
- {
+ variable sized operations. For task reductions we use 4 passes, in the
+ first two we ignore them, in the third one gather arguments for
+ GOMP_task_reduction_remap call and in the last pass actually handle
+ the task reductions. */
+ for (pass = 0; pass < ((task_reduction_cnt || task_reduction_other_cnt)
+ ? 4 : 2); ++pass)
+ {
+ if (pass == 2 && task_reduction_cnt)
+ {
+ tskred_atype
+ = build_array_type_nelts (ptr_type_node, task_reduction_cnt
+ + task_reduction_cntorig);
+ tskred_avar = create_tmp_var_raw (tskred_atype);
+ gimple_add_tmp_var (tskred_avar);
+ TREE_ADDRESSABLE (tskred_avar) = 1;
+ task_reduction_cnt_full = task_reduction_cnt;
+ task_reduction_cntorig_full = task_reduction_cntorig;
+ }
+ else if (pass == 3 && task_reduction_cnt)
+ {
+ x = builtin_decl_explicit (BUILT_IN_GOMP_TASK_REDUCTION_REMAP);
+ gimple *g
+ = gimple_build_call (x, 3, size_int (task_reduction_cnt),
+ size_int (task_reduction_cntorig),
+ build_fold_addr_expr (tskred_avar));
+ gimple_seq_add_stmt (ilist, g);
+ }
+ if (pass == 3 && task_reduction_other_cnt)
+ {
+ /* For reduction clauses, build
+ tskred_base = (void *) tskred_temp[2]
+ + omp_get_thread_num () * tskred_temp[1]
+ or if tskred_temp[1] is known to be constant, that constant
+ directly. This is the start of the private reduction copy block
+ for the current thread. */
+ tree v = create_tmp_var (integer_type_node);
+ x = builtin_decl_explicit (BUILT_IN_OMP_GET_THREAD_NUM);
+ gimple *g = gimple_build_call (x, 0);
+ gimple_call_set_lhs (g, v);
+ gimple_seq_add_stmt (ilist, g);
+ c = omp_find_clause (clauses, OMP_CLAUSE__REDUCTEMP_);
+ tskred_temp = OMP_CLAUSE_DECL (c);
+ if (is_taskreg_ctx (ctx))
+ tskred_temp = lookup_decl (tskred_temp, ctx);
+ tree v2 = create_tmp_var (sizetype);
+ g = gimple_build_assign (v2, NOP_EXPR, v);
+ gimple_seq_add_stmt (ilist, g);
+ if (ctx->task_reductions[0])
+ v = fold_convert (sizetype, ctx->task_reductions[0]);
+ else
+ v = task_reduction_read (ilist, tskred_temp, sizetype, 1);
+ tree v3 = create_tmp_var (sizetype);
+ g = gimple_build_assign (v3, MULT_EXPR, v2, v);
+ gimple_seq_add_stmt (ilist, g);
+ v = task_reduction_read (ilist, tskred_temp, ptr_type_node, 2);
+ tskred_base = create_tmp_var (ptr_type_node);
+ g = gimple_build_assign (tskred_base, POINTER_PLUS_EXPR, v, v3);
+ gimple_seq_add_stmt (ilist, g);
+ }
+ task_reduction_cnt = 0;
+ task_reduction_cntorig = 0;
+ task_reduction_other_cnt = 0;
for (c = clauses; c ; c = OMP_CLAUSE_CHAIN (c))
{
enum omp_clause_code c_kind = OMP_CLAUSE_CODE (c);
tree var, new_var;
bool by_ref;
location_t clause_loc = OMP_CLAUSE_LOCATION (c);
+ bool task_reduction_p = false;
+ bool task_reduction_needs_orig_p = false;
+ tree cond = NULL_TREE;
switch (c_kind)
{
@@ -3590,8 +3892,10 @@ lower_rec_input_clauses (tree clauses, g
continue;
break;
case OMP_CLAUSE_SHARED:
- /* Ignore shared directives in teams construct. */
- if (gimple_code (ctx->stmt) == GIMPLE_OMP_TEAMS)
+ /* Ignore shared directives in teams construct inside
+ of target construct. */
+ if (gimple_code (ctx->stmt) == GIMPLE_OMP_TEAMS
+ && !is_host_teams_ctx (ctx))
continue;
if (maybe_lookup_decl (OMP_CLAUSE_DECL (c), ctx) == NULL)
{
@@ -3608,11 +3912,46 @@ lower_rec_input_clauses (tree clauses, g
lastprivate_firstprivate = true;
break;
case OMP_CLAUSE_REDUCTION:
- if (OMP_CLAUSE_REDUCTION_OMP_ORIG_REF (c))
+ case OMP_CLAUSE_IN_REDUCTION:
+ if (is_task_ctx (ctx) || OMP_CLAUSE_REDUCTION_TASK (c))
+ {
+ task_reduction_p = true;
+ if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_REDUCTION)
+ {
+ task_reduction_other_cnt++;
+ if (pass == 2)
+ continue;
+ }
+ else
+ task_reduction_cnt++;
+ if (OMP_CLAUSE_REDUCTION_OMP_ORIG_REF (c))
+ {
+ var = OMP_CLAUSE_DECL (c);
+ /* If var is a global variable that isn't privatized
+ in outer contexts, we don't need to look up the
+ original address, it is always the address of the
+ global variable itself. */
+ if (!DECL_P (var)
+ || omp_is_reference (var)
+ || !is_global_var
+ (maybe_lookup_decl_in_outer_ctx (var, ctx)))
+ {
+ task_reduction_needs_orig_p = true;
+ if (OMP_CLAUSE_CODE (c) != OMP_CLAUSE_REDUCTION)
+ task_reduction_cntorig++;
+ }
+ }
+ }
+ else if (OMP_CLAUSE_REDUCTION_OMP_ORIG_REF (c))
reduction_omp_orig_ref = true;
break;
+ case OMP_CLAUSE__REDUCTEMP_:
+ if (!is_taskreg_ctx (ctx))
+ continue;
+ /* FALLTHRU */
case OMP_CLAUSE__LOOPTEMP_:
- /* Handle _looptemp_ clauses only on parallel/task. */
+ /* Handle _looptemp_/_reductemp_ clauses only on
+ parallel/task. */
if (fd)
continue;
break;
@@ -3632,7 +3971,7 @@ lower_rec_input_clauses (tree clauses, g
lastprivate_firstprivate = true;
break;
case OMP_CLAUSE_ALIGNED:
- if (pass == 0)
+ if (pass != 1)
continue;
var = OMP_CLAUSE_DECL (c);
if (TREE_CODE (TREE_TYPE (var)) == POINTER_TYPE
@@ -3673,8 +4012,13 @@ lower_rec_input_clauses (tree clauses, g
continue;
}
+ if (task_reduction_p != (pass >= 2))
+ continue;
+
new_var = var = OMP_CLAUSE_DECL (c);
- if (c_kind == OMP_CLAUSE_REDUCTION && TREE_CODE (var) == MEM_REF)
+ if ((c_kind == OMP_CLAUSE_REDUCTION
+ || c_kind == OMP_CLAUSE_IN_REDUCTION)
+ && TREE_CODE (var) == MEM_REF)
{
var = TREE_OPERAND (var, 0);
if (TREE_CODE (var) == POINTER_PLUS_EXPR)
@@ -3701,7 +4045,8 @@ lower_rec_input_clauses (tree clauses, g
continue;
}
/* C/C++ array section reductions. */
- else if (c_kind == OMP_CLAUSE_REDUCTION
+ else if ((c_kind == OMP_CLAUSE_REDUCTION
+ || c_kind == OMP_CLAUSE_IN_REDUCTION)
&& var != OMP_CLAUSE_DECL (c))
{
if (pass == 0)
@@ -3709,6 +4054,7 @@ lower_rec_input_clauses (tree clauses, g
tree bias = TREE_OPERAND (OMP_CLAUSE_DECL (c), 1);
tree orig_var = TREE_OPERAND (OMP_CLAUSE_DECL (c), 0);
+
if (TREE_CODE (orig_var) == POINTER_PLUS_EXPR)
{
tree b = TREE_OPERAND (orig_var, 1);
@@ -3729,6 +4075,47 @@ lower_rec_input_clauses (tree clauses, g
}
orig_var = TREE_OPERAND (orig_var, 0);
}
+ if (pass == 2)
+ {
+ tree out = maybe_lookup_decl_in_outer_ctx (var, ctx);
+ if (is_global_var (out)
+ && TREE_CODE (TREE_TYPE (out)) != POINTER_TYPE
+ && (TREE_CODE (TREE_TYPE (out)) != REFERENCE_TYPE
+ || (TREE_CODE (TREE_TYPE (TREE_TYPE (out)))
+ != POINTER_TYPE)))
+ x = var;
+ else
+ {
+ bool by_ref = use_pointer_for_field (var, NULL);
+ x = build_receiver_ref (var, by_ref, ctx);
+ if (TREE_CODE (TREE_TYPE (var)) == REFERENCE_TYPE
+ && (TREE_CODE (TREE_TYPE (TREE_TYPE (var)))
+ == POINTER_TYPE))
+ x = build_fold_addr_expr (x);
+ }
+ if (TREE_CODE (orig_var) == INDIRECT_REF)
+ x = build_simple_mem_ref (x);
+ else if (TREE_CODE (orig_var) == ADDR_EXPR)
+ {
+ if (var == TREE_OPERAND (orig_var, 0))
+ x = build_fold_addr_expr (x);
+ }
+ bias = fold_convert (sizetype, bias);
+ x = fold_convert (ptr_type_node, x);
+ x = fold_build2_loc (clause_loc, POINTER_PLUS_EXPR,
+ TREE_TYPE (x), x, bias);
+ unsigned cnt = task_reduction_cnt - 1;
+ if (!task_reduction_needs_orig_p)
+ cnt += (task_reduction_cntorig_full
+ - task_reduction_cntorig);
+ else
+ cnt = task_reduction_cntorig - 1;
+ tree r = build4 (ARRAY_REF, ptr_type_node, tskred_avar,
+ size_int (cnt), NULL_TREE, NULL_TREE);
+ gimplify_assign (r, x, ilist);
+ continue;
+ }
+
if (TREE_CODE (orig_var) == INDIRECT_REF
|| TREE_CODE (orig_var) == ADDR_EXPR)
orig_var = TREE_OPERAND (orig_var, 0);
@@ -3737,7 +4124,64 @@ lower_rec_input_clauses (tree clauses, g
gcc_assert (TREE_CODE (type) == ARRAY_TYPE);
tree v = TYPE_MAX_VALUE (TYPE_DOMAIN (type));
const char *name = get_name (orig_var);
- if (TREE_CONSTANT (v))
+ if (pass == 3)
+ {
+ tree xv = create_tmp_var (ptr_type_node);
+ if (OMP_CLAUSE_CODE (c) != OMP_CLAUSE_REDUCTION)
+ {
+ unsigned cnt = task_reduction_cnt - 1;
+ if (!task_reduction_needs_orig_p)
+ cnt += (task_reduction_cntorig_full
+ - task_reduction_cntorig);
+ else
+ cnt = task_reduction_cntorig - 1;
+ x = build4 (ARRAY_REF, ptr_type_node, tskred_avar,
+ size_int (cnt), NULL_TREE, NULL_TREE);
+
+ gimple *g = gimple_build_assign (xv, x);
+ gimple_seq_add_stmt (ilist, g);
+ }
+ else
+ {
+ unsigned int idx = *ctx->task_reduction_map->get (c);
+ tree off;
+ if (ctx->task_reductions[1 + idx])
+ off = fold_convert (sizetype,
+ ctx->task_reductions[1 + idx]);
+ else
+ off = task_reduction_read (ilist, tskred_temp, sizetype,
+ 7 + 3 * idx + 1);
+ gimple *g = gimple_build_assign (xv, POINTER_PLUS_EXPR,
+ tskred_base, off);
+ gimple_seq_add_stmt (ilist, g);
+ }
+ x = fold_convert (build_pointer_type (boolean_type_node),
+ xv);
+ if (TREE_CONSTANT (v))
+ x = fold_build2 (POINTER_PLUS_EXPR, TREE_TYPE (x), x,
+ TYPE_SIZE_UNIT (type));
+ else
+ {
+ tree t = maybe_lookup_decl (v, ctx);
+ if (t)
+ v = t;
+ else
+ v = maybe_lookup_decl_in_outer_ctx (v, ctx);
+ gimplify_expr (&v, ilist, NULL, is_gimple_val,
+ fb_rvalue);
+ t = fold_build2_loc (clause_loc, PLUS_EXPR,
+ TREE_TYPE (v), v,
+ build_int_cst (TREE_TYPE (v), 1));
+ t = fold_build2_loc (clause_loc, MULT_EXPR,
+ TREE_TYPE (v), t,
+ TYPE_SIZE_UNIT (TREE_TYPE (type)));
+ x = fold_build2 (POINTER_PLUS_EXPR, TREE_TYPE (x), x, t);
+ }
+ cond = create_tmp_var (TREE_TYPE (x));
+ gimplify_assign (cond, x, ilist);
+ x = xv;
+ }
+ else if (TREE_CONSTANT (v))
{
x = create_tmp_var_raw (type, name);
gimple_add_tmp_var (x);
@@ -3799,7 +4243,7 @@ lower_rec_input_clauses (tree clauses, g
tree new_orig_var = lookup_decl (orig_var, ctx);
tree t = build_fold_indirect_ref (new_var);
DECL_IGNORED_P (new_var) = 0;
- TREE_THIS_NOTRAP (t);
+ TREE_THIS_NOTRAP (t) = 1;
SET_DECL_VALUE_EXPR (new_orig_var, t);
DECL_HAS_VALUE_EXPR_P (new_orig_var) = 1;
}
@@ -3824,44 +4268,101 @@ lower_rec_input_clauses (tree clauses, g
x = fold_convert_loc (clause_loc, TREE_TYPE (new_var), x);
gimplify_assign (new_var, x, ilist);
}
- tree y1 = create_tmp_var (ptype, NULL);
+ /* GOMP_taskgroup_reduction_register memsets the whole
+ array to zero. If the initializer is zero, we don't
+ need to initialize it again, just mark it as ever
+ used unconditionally, i.e. cond = true. */
+ if (cond
+ && OMP_CLAUSE_REDUCTION_PLACEHOLDER (c) == NULL_TREE
+ && initializer_zerop (omp_reduction_init (c,
+ TREE_TYPE (type))))
+ {
+ gimple *g = gimple_build_assign (build_simple_mem_ref (cond),
+ boolean_true_node);
+ gimple_seq_add_stmt (ilist, g);
+ continue;
+ }
+ tree end = create_artificial_label (UNKNOWN_LOCATION);
+ if (cond)
+ {
+ gimple *g;
+ if (!is_parallel_ctx (ctx))
+ {
+ tree condv = create_tmp_var (boolean_type_node);
+ g = gimple_build_assign (condv,
+ build_simple_mem_ref (cond));
+ gimple_seq_add_stmt (ilist, g);
+ tree lab1 = create_artificial_label (UNKNOWN_LOCATION);
+ g = gimple_build_cond (NE_EXPR, condv,
+ boolean_false_node, end, lab1);
+ gimple_seq_add_stmt (ilist, g);
+ gimple_seq_add_stmt (ilist, gimple_build_label (lab1));
+ }
+ g = gimple_build_assign (build_simple_mem_ref (cond),
+ boolean_true_node);
+ gimple_seq_add_stmt (ilist, g);
+ }
+
+ tree y1 = create_tmp_var (ptype);
gimplify_assign (y1, y, ilist);
tree i2 = NULL_TREE, y2 = NULL_TREE;
tree body2 = NULL_TREE, end2 = NULL_TREE;
tree y3 = NULL_TREE, y4 = NULL_TREE;
- if (OMP_CLAUSE_REDUCTION_PLACEHOLDER (c) || is_simd)
+ if (task_reduction_needs_orig_p)
{
- y2 = create_tmp_var (ptype, NULL);
- gimplify_assign (y2, y, ilist);
- tree ref = build_outer_var_ref (var, ctx);
- /* For ref build_outer_var_ref already performs this. */
- if (TREE_CODE (d) == INDIRECT_REF)
- gcc_assert (omp_is_reference (var));
- else if (TREE_CODE (d) == ADDR_EXPR)
- ref = build_fold_addr_expr (ref);
- else if (omp_is_reference (var))
- ref = build_fold_addr_expr (ref);
- ref = fold_convert_loc (clause_loc, ptype, ref);
- if (OMP_CLAUSE_REDUCTION_PLACEHOLDER (c)
- && OMP_CLAUSE_REDUCTION_OMP_ORIG_REF (c))
+ y3 = create_tmp_var (ptype);
+ tree ref;
+ if (OMP_CLAUSE_CODE (c) != OMP_CLAUSE_REDUCTION)
+ ref = build4 (ARRAY_REF, ptr_type_node, tskred_avar,
+ size_int (task_reduction_cnt_full
+ + task_reduction_cntorig - 1),
+ NULL_TREE, NULL_TREE);
+ else
{
- y3 = create_tmp_var (ptype, NULL);
- gimplify_assign (y3, unshare_expr (ref), ilist);
+ unsigned int idx = *ctx->task_reduction_map->get (c);
+ ref = task_reduction_read (ilist, tskred_temp, ptype,
+ 7 + 3 * idx);
}
- if (is_simd)
+ gimplify_assign (y3, ref, ilist);
+ }
+ else if (OMP_CLAUSE_REDUCTION_PLACEHOLDER (c) || is_simd)
+ {
+ if (pass != 3)
{
- y4 = create_tmp_var (ptype, NULL);
- gimplify_assign (y4, ref, dlist);
+ y2 = create_tmp_var (ptype);
+ gimplify_assign (y2, y, ilist);
+ }
+ if (is_simd || OMP_CLAUSE_REDUCTION_OMP_ORIG_REF (c))
+ {
+ tree ref = build_outer_var_ref (var, ctx);
+ /* For ref build_outer_var_ref already performs this. */
+ if (TREE_CODE (d) == INDIRECT_REF)
+ gcc_assert (omp_is_reference (var));
+ else if (TREE_CODE (d) == ADDR_EXPR)
+ ref = build_fold_addr_expr (ref);
+ else if (omp_is_reference (var))
+ ref = build_fold_addr_expr (ref);
+ ref = fold_convert_loc (clause_loc, ptype, ref);
+ if (OMP_CLAUSE_REDUCTION_PLACEHOLDER (c)
+ && OMP_CLAUSE_REDUCTION_OMP_ORIG_REF (c))
+ {
+ y3 = create_tmp_var (ptype);
+ gimplify_assign (y3, unshare_expr (ref), ilist);
+ }
+ if (is_simd)
+ {
+ y4 = create_tmp_var (ptype);
+ gimplify_assign (y4, ref, dlist);
+ }
}
}
- tree i = create_tmp_var (TREE_TYPE (v), NULL);
+ tree i = create_tmp_var (TREE_TYPE (v));
gimplify_assign (i, build_int_cst (TREE_TYPE (v), 0), ilist);
tree body = create_artificial_label (UNKNOWN_LOCATION);
- tree end = create_artificial_label (UNKNOWN_LOCATION);
gimple_seq_add_stmt (ilist, gimple_build_label (body));
if (y2)
{
- i2 = create_tmp_var (TREE_TYPE (v), NULL);
+ i2 = create_tmp_var (TREE_TYPE (v));
gimplify_assign (i2, build_int_cst (TREE_TYPE (v), 0), dlist);
body2 = create_artificial_label (UNKNOWN_LOCATION);
end2 = create_artificial_label (UNKNOWN_LOCATION);
@@ -3904,14 +4405,17 @@ lower_rec_input_clauses (tree clauses, g
}
DECL_HAS_VALUE_EXPR_P (placeholder) = 0;
DECL_HAS_VALUE_EXPR_P (decl_placeholder) = 0;
- x = lang_hooks.decls.omp_clause_dtor
- (c, build_simple_mem_ref (y2));
- if (x)
+ if (y2)
{
- gimple_seq tseq = NULL;
- dtor = x;
- gimplify_stmt (&dtor, &tseq);
- gimple_seq_add_seq (dlist, tseq);
+ x = lang_hooks.decls.omp_clause_dtor
+ (c, build_simple_mem_ref (y2));
+ if (x)
+ {
+ gimple_seq tseq = NULL;
+ dtor = x;
+ gimplify_stmt (&dtor, &tseq);
+ gimple_seq_add_seq (dlist, tseq);
+ }
}
}
else
@@ -3970,6 +4474,78 @@ lower_rec_input_clauses (tree clauses, g
}
continue;
}
+ else if (pass == 2)
+ {
+ if (is_global_var (maybe_lookup_decl_in_outer_ctx (var, ctx)))
+ x = var;
+ else
+ {
+ bool by_ref = use_pointer_for_field (var, ctx);
+ x = build_receiver_ref (var, by_ref, ctx);
+ }
+ if (!omp_is_reference (var))
+ x = build_fold_addr_expr (x);
+ x = fold_convert (ptr_type_node, x);
+ unsigned cnt = task_reduction_cnt - 1;
+ if (!task_reduction_needs_orig_p)
+ cnt += task_reduction_cntorig_full - task_reduction_cntorig;
+ else
+ cnt = task_reduction_cntorig - 1;
+ tree r = build4 (ARRAY_REF, ptr_type_node, tskred_avar,
+ size_int (cnt), NULL_TREE, NULL_TREE);
+ gimplify_assign (r, x, ilist);
+ continue;
+ }
+ else if (pass == 3)
+ {
+ tree type = TREE_TYPE (new_var);
+ if (!omp_is_reference (var))
+ type = build_pointer_type (type);
+ if (OMP_CLAUSE_CODE (c) != OMP_CLAUSE_REDUCTION)
+ {
+ unsigned cnt = task_reduction_cnt - 1;
+ if (!task_reduction_needs_orig_p)
+ cnt += (task_reduction_cntorig_full
+ - task_reduction_cntorig);
+ else
+ cnt = task_reduction_cntorig - 1;
+ x = build4 (ARRAY_REF, ptr_type_node, tskred_avar,
+ size_int (cnt), NULL_TREE, NULL_TREE);
+ }
+ else
+ {
+ unsigned int idx = *ctx->task_reduction_map->get (c);
+ tree off;
+ if (ctx->task_reductions[1 + idx])
+ off = fold_convert (sizetype,
+ ctx->task_reductions[1 + idx]);
+ else
+ off = task_reduction_read (ilist, tskred_temp, sizetype,
+ 7 + 3 * idx + 1);
+ x = fold_build2 (POINTER_PLUS_EXPR, ptr_type_node,
+ tskred_base, off);
+ }
+ x = fold_convert (type, x);
+ tree t;
+ if (omp_is_reference (var))
+ {
+ gimplify_assign (new_var, x, ilist);
+ t = new_var;
+ new_var = build_simple_mem_ref (new_var);
+ }
+ else
+ {
+ t = create_tmp_var (type);
+ gimplify_assign (t, x, ilist);
+ SET_DECL_VALUE_EXPR (new_var, build_simple_mem_ref (t));
+ DECL_HAS_VALUE_EXPR_P (new_var) = 1;
+ }
+ t = fold_convert (build_pointer_type (boolean_type_node), t);
+ t = fold_build2 (POINTER_PLUS_EXPR, TREE_TYPE (t), t,
+ TYPE_SIZE_UNIT (TREE_TYPE (type)));
+ cond = create_tmp_var (TREE_TYPE (t));
+ gimplify_assign (cond, t, ilist);
+ }
else if (is_variable_sized (var))
{
/* For variable sized types, we need to allocate the
@@ -4003,7 +4579,9 @@ lower_rec_input_clauses (tree clauses, g
gimplify_assign (ptr, x, ilist);
}
}
- else if (omp_is_reference (var))
+ else if (omp_is_reference (var)
+ && (c_kind != OMP_CLAUSE_FIRSTPRIVATE
+ || !OMP_CLAUSE_FIRSTPRIVATE_NO_REFERENCE (c)))
{
/* For references that are being privatized for Fortran,
allocate new backing storage for the new pointer
@@ -4053,7 +4631,8 @@ lower_rec_input_clauses (tree clauses, g
new_var = build_simple_mem_ref_loc (clause_loc, new_var);
}
- else if (c_kind == OMP_CLAUSE_REDUCTION
+ else if ((c_kind == OMP_CLAUSE_REDUCTION
+ || c_kind == OMP_CLAUSE_IN_REDUCTION)
&& OMP_CLAUSE_REDUCTION_PLACEHOLDER (c))
{
if (pass == 0)
@@ -4065,8 +4644,10 @@ lower_rec_input_clauses (tree clauses, g
switch (OMP_CLAUSE_CODE (c))
{
case OMP_CLAUSE_SHARED:
- /* Ignore shared directives in teams construct. */
- if (gimple_code (ctx->stmt) == GIMPLE_OMP_TEAMS)
+ /* Ignore shared directives in teams construct inside
+ target construct. */
+ if (gimple_code (ctx->stmt) == GIMPLE_OMP_TEAMS
+ && !is_host_teams_ctx (ctx))
continue;
/* Shared global vars are just accessed directly. */
if (is_global_var (new_var))
@@ -4170,7 +4751,9 @@ lower_rec_input_clauses (tree clauses, g
case OMP_CLAUSE_FIRSTPRIVATE:
if (is_task_ctx (ctx))
{
- if (omp_is_reference (var) || is_variable_sized (var))
+ if ((omp_is_reference (var)
+ && !OMP_CLAUSE_FIRSTPRIVATE_NO_REFERENCE (c))
+ || is_variable_sized (var))
goto do_dtor;
else if (is_global_var (maybe_lookup_decl_in_outer_ctx (var,
ctx))
@@ -4182,6 +4765,18 @@ lower_rec_input_clauses (tree clauses, g
goto do_dtor;
}
}
+ if (OMP_CLAUSE_FIRSTPRIVATE_NO_REFERENCE (c)
+ && omp_is_reference (var))
+ {
+ x = build_outer_var_ref (var, ctx);
+ gcc_assert (TREE_CODE (x) == MEM_REF
+ && integer_zerop (TREE_OPERAND (x, 1)));
+ x = TREE_OPERAND (x, 0);
+ x = lang_hooks.decls.omp_clause_copy_ctor
+ (c, unshare_expr (new_var), x);
+ gimplify_and_add (x, ilist);
+ goto do_dtor;
+ }
do_firstprivate:
x = build_outer_var_ref (var, ctx);
if (is_simd)
@@ -4273,6 +4868,7 @@ lower_rec_input_clauses (tree clauses, g
goto do_dtor;
case OMP_CLAUSE__LOOPTEMP_:
+ case OMP_CLAUSE__REDUCTEMP_:
gcc_assert (is_taskreg_ctx (ctx));
x = build_outer_var_ref (var, ctx);
x = build2 (MODIFY_EXPR, TREE_TYPE (new_var), new_var, x);
@@ -4288,6 +4884,7 @@ lower_rec_input_clauses (tree clauses, g
break;
case OMP_CLAUSE_REDUCTION:
+ case OMP_CLAUSE_IN_REDUCTION:
/* OpenACC reductions are initialized using the
GOACC_REDUCTION internal function. */
if (is_gimple_omp_oacc (ctx->stmt))
@@ -4296,12 +4893,40 @@ lower_rec_input_clauses (tree clauses, g
{
tree placeholder = OMP_CLAUSE_REDUCTION_PLACEHOLDER (c);
gimple *tseq;
- x = build_outer_var_ref (var, ctx);
+ tree ptype = TREE_TYPE (placeholder);
+ if (cond)
+ {
+ x = error_mark_node;
+ if (OMP_CLAUSE_REDUCTION_OMP_ORIG_REF (c)
+ && !task_reduction_needs_orig_p)
+ x = var;
+ else if (OMP_CLAUSE_REDUCTION_OMP_ORIG_REF (c))
+ {
+ tree pptype = build_pointer_type (ptype);
+ if (OMP_CLAUSE_CODE (c) != OMP_CLAUSE_REDUCTION)
+ x = build4 (ARRAY_REF, ptr_type_node, tskred_avar,
+ size_int (task_reduction_cnt_full
+ + task_reduction_cntorig - 1),
+ NULL_TREE, NULL_TREE);
+ else
+ {
+ unsigned int idx
+ = *ctx->task_reduction_map->get (c);
+ x = task_reduction_read (ilist, tskred_temp,
+ pptype, 7 + 3 * idx);
+ }
+ x = fold_convert (pptype, x);
+ x = build_simple_mem_ref (x);
+ }
+ }
+ else
+ {
+ x = build_outer_var_ref (var, ctx);
- if (omp_is_reference (var)
- && !useless_type_conversion_p (TREE_TYPE (placeholder),
- TREE_TYPE (x)))
- x = build_fold_addr_expr_loc (clause_loc, x);
+ if (omp_is_reference (var)
+ && !useless_type_conversion_p (ptype, TREE_TYPE (x)))
+ x = build_fold_addr_expr_loc (clause_loc, x);
+ }
SET_DECL_VALUE_EXPR (placeholder, x);
DECL_HAS_VALUE_EXPR_P (placeholder) = 1;
tree new_vard = new_var;
@@ -4365,9 +4990,35 @@ lower_rec_input_clauses (tree clauses, g
initialization now. */
else if (omp_is_reference (var) && is_simd)
handle_simd_reference (clause_loc, new_vard, ilist);
+
+ tree lab2 = NULL_TREE;
+ if (cond)
+ {
+ gimple *g;
+ if (!is_parallel_ctx (ctx))
+ {
+ tree condv = create_tmp_var (boolean_type_node);
+ tree m = build_simple_mem_ref (cond);
+ g = gimple_build_assign (condv, m);
+ gimple_seq_add_stmt (ilist, g);
+ tree lab1
+ = create_artificial_label (UNKNOWN_LOCATION);
+ lab2 = create_artificial_label (UNKNOWN_LOCATION);
+ g = gimple_build_cond (NE_EXPR, condv,
+ boolean_false_node,
+ lab2, lab1);
+ gimple_seq_add_stmt (ilist, g);
+ gimple_seq_add_stmt (ilist,
+ gimple_build_label (lab1));
+ }
+ g = gimple_build_assign (build_simple_mem_ref (cond),
+ boolean_true_node);
+ gimple_seq_add_stmt (ilist, g);
+ }
x = lang_hooks.decls.omp_clause_default_ctor
(c, unshare_expr (new_var),
- build_outer_var_ref (var, ctx));
+ cond ? NULL_TREE
+ : build_outer_var_ref (var, ctx));
if (x)
gimplify_and_add (x, ilist);
if (OMP_CLAUSE_REDUCTION_GIMPLE_INIT (c))
@@ -4385,6 +5036,12 @@ lower_rec_input_clauses (tree clauses, g
OMP_CLAUSE_REDUCTION_GIMPLE_MERGE (c) = NULL;
}
DECL_HAS_VALUE_EXPR_P (placeholder) = 0;
+ if (cond)
+ {
+ if (lab2)
+ gimple_seq_add_stmt (ilist, gimple_build_label (lab2));
+ break;
+ }
goto do_dtor;
}
else
@@ -4393,6 +5050,49 @@ lower_rec_input_clauses (tree clauses, g
gcc_assert (TREE_CODE (TREE_TYPE (new_var)) != ARRAY_TYPE);
enum tree_code code = OMP_CLAUSE_REDUCTION_CODE (c);
+ if (cond)
+ {
+ gimple *g;
+ tree lab2 = NULL_TREE;
+ /* GOMP_taskgroup_reduction_register memsets the whole
+ array to zero. If the initializer is zero, we don't
+ need to initialize it again, just mark it as ever
+ used unconditionally, i.e. cond = true. */
+ if (initializer_zerop (x))
+ {
+ g = gimple_build_assign (build_simple_mem_ref (cond),
+ boolean_true_node);
+ gimple_seq_add_stmt (ilist, g);
+ break;
+ }
+
+ /* Otherwise, emit
+ if (!cond) { cond = true; new_var = x; } */
+ if (!is_parallel_ctx (ctx))
+ {
+ tree condv = create_tmp_var (boolean_type_node);
+ tree m = build_simple_mem_ref (cond);
+ g = gimple_build_assign (condv, m);
+ gimple_seq_add_stmt (ilist, g);
+ tree lab1
+ = create_artificial_label (UNKNOWN_LOCATION);
+ lab2 = create_artificial_label (UNKNOWN_LOCATION);
+ g = gimple_build_cond (NE_EXPR, condv,
+ boolean_false_node,
+ lab2, lab1);
+ gimple_seq_add_stmt (ilist, g);
+ gimple_seq_add_stmt (ilist,
+ gimple_build_label (lab1));
+ }
+ g = gimple_build_assign (build_simple_mem_ref (cond),
+ boolean_true_node);
+ gimple_seq_add_stmt (ilist, g);
+ gimplify_assign (new_var, x, ilist);
+ if (lab2)
+ gimple_seq_add_stmt (ilist, gimple_build_label (lab2));
+ break;
+ }
+
/* reduction(-:var) sums up the partial results, so it
acts identically to reduction(+:var). */
if (code == MINUS_EXPR)
@@ -4456,6 +5156,12 @@ lower_rec_input_clauses (tree clauses, g
}
}
}
+ if (tskred_avar)
+ {
+ tree clobber = build_constructor (TREE_TYPE (tskred_avar), NULL);
+ TREE_THIS_VOLATILE (clobber) = 1;
+ gimple_seq_add_stmt (ilist, gimple_build_assign (tskred_avar, clobber));
+ }
if (known_eq (sctx.max_vf, 1U))
sctx.is_simt = false;
@@ -4587,8 +5293,9 @@ lower_rec_input_clauses (tree clauses, g
{
/* Don't add any barrier for #pragma omp simd or
#pragma omp distribute. */
- if (gimple_code (ctx->stmt) != GIMPLE_OMP_FOR
- || gimple_omp_for_kind (ctx->stmt) == GF_OMP_FOR_KIND_FOR)
+ if (!is_task_ctx (ctx)
+ && (gimple_code (ctx->stmt) != GIMPLE_OMP_FOR
+ || gimple_omp_for_kind (ctx->stmt) == GF_OMP_FOR_KIND_FOR))
gimple_seq_add_stmt (ilist, omp_build_barrier (NULL_TREE));
}
@@ -5078,7 +5785,8 @@ lower_reduction_clauses (tree clauses, g
/* First see if there is exactly one reduction clause. Use OMP_ATOMIC
update in that case, otherwise use a lock. */
for (c = clauses; c && count < 2; c = OMP_CLAUSE_CHAIN (c))
- if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_REDUCTION)
+ if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_REDUCTION
+ && !OMP_CLAUSE_REDUCTION_TASK (c))
{
if (OMP_CLAUSE_REDUCTION_PLACEHOLDER (c)
|| TREE_CODE (OMP_CLAUSE_DECL (c)) == MEM_REF)
@@ -5099,7 +5807,8 @@ lower_reduction_clauses (tree clauses, g
enum tree_code code;
location_t clause_loc = OMP_CLAUSE_LOCATION (c);
- if (OMP_CLAUSE_CODE (c) != OMP_CLAUSE_REDUCTION)
+ if (OMP_CLAUSE_CODE (c) != OMP_CLAUSE_REDUCTION
+ || OMP_CLAUSE_REDUCTION_TASK (c))
continue;
enum omp_clause_code ccode = OMP_CLAUSE_REDUCTION;
@@ -5150,6 +5859,7 @@ lower_reduction_clauses (tree clauses, g
ref = build1 (INDIRECT_REF, TREE_TYPE (TREE_TYPE (addr)), addr);
x = fold_build2_loc (clause_loc, code, TREE_TYPE (ref), ref, new_var);
x = build2 (OMP_ATOMIC, void_type_node, addr, x);
+ OMP_ATOMIC_MEMORY_ORDER (x) = OMP_MEMORY_ORDER_RELAXED;
gimplify_and_add (x, stmt_seqp);
return;
}
@@ -5158,7 +5868,7 @@ lower_reduction_clauses (tree clauses, g
tree d = OMP_CLAUSE_DECL (c);
tree type = TREE_TYPE (d);
tree v = TYPE_MAX_VALUE (TYPE_DOMAIN (type));
- tree i = create_tmp_var (TREE_TYPE (v), NULL);
+ tree i = create_tmp_var (TREE_TYPE (v));
tree ptype = build_pointer_type (TREE_TYPE (type));
tree bias = TREE_OPERAND (d, 1);
d = TREE_OPERAND (d, 0);
@@ -5222,10 +5932,10 @@ lower_reduction_clauses (tree clauses, g
}
new_var = fold_convert_loc (clause_loc, ptype, new_var);
ref = fold_convert_loc (clause_loc, ptype, ref);
- tree m = create_tmp_var (ptype, NULL);
+ tree m = create_tmp_var (ptype);
gimplify_assign (m, new_var, stmt_seqp);
new_var = m;
- m = create_tmp_var (ptype, NULL);
+ m = create_tmp_var (ptype);
gimplify_assign (m, ref, stmt_seqp);
ref = m;
gimplify_assign (i, build_int_cst (TREE_TYPE (v), 0), stmt_seqp);
@@ -5387,7 +6097,12 @@ lower_send_clauses (tree clauses, gimple
case OMP_CLAUSE_FIRSTPRIVATE:
case OMP_CLAUSE_COPYIN:
case OMP_CLAUSE_LASTPRIVATE:
+ case OMP_CLAUSE_IN_REDUCTION:
+ case OMP_CLAUSE__REDUCTEMP_:
+ break;
case OMP_CLAUSE_REDUCTION:
+ if (is_task_ctx (ctx) || OMP_CLAUSE_REDUCTION_TASK (c))
+ continue;
break;
case OMP_CLAUSE_SHARED:
if (OMP_CLAUSE_SHARED_FIRSTPRIVATE (c))
@@ -5405,7 +6120,8 @@ lower_send_clauses (tree clauses, gimple
}
val = OMP_CLAUSE_DECL (c);
- if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_REDUCTION
+ if ((OMP_CLAUSE_CODE (c) == OMP_CLAUSE_REDUCTION
+ || OMP_CLAUSE_CODE (c) == OMP_CLAUSE_IN_REDUCTION)
&& TREE_CODE (val) == MEM_REF)
{
val = TREE_OPERAND (val, 0);
@@ -5429,7 +6145,13 @@ lower_send_clauses (tree clauses, gimple
var = lookup_decl_in_outer_ctx (val, ctx_for_o);
if (OMP_CLAUSE_CODE (c) != OMP_CLAUSE_COPYIN
- && is_global_var (var))
+ && is_global_var (var)
+ && (val == OMP_CLAUSE_DECL (c)
+ || !is_task_ctx (ctx)
+ || (TREE_CODE (TREE_TYPE (val)) != POINTER_TYPE
+ && (TREE_CODE (TREE_TYPE (val)) != REFERENCE_TYPE
+ || (TREE_CODE (TREE_TYPE (TREE_TYPE (val)))
+ != POINTER_TYPE)))))
continue;
t = omp_member_access_dummy_var (var);
@@ -5457,7 +6179,8 @@ lower_send_clauses (tree clauses, gimple
continue;
}
- if ((OMP_CLAUSE_CODE (c) != OMP_CLAUSE_REDUCTION
+ if (((OMP_CLAUSE_CODE (c) != OMP_CLAUSE_REDUCTION
+ && OMP_CLAUSE_CODE (c) != OMP_CLAUSE_IN_REDUCTION)
|| val == OMP_CLAUSE_DECL (c))
&& is_variable_sized (val))
continue;
@@ -5476,6 +6199,7 @@ lower_send_clauses (tree clauses, gimple
case OMP_CLAUSE_PRIVATE:
case OMP_CLAUSE_COPYIN:
case OMP_CLAUSE__LOOPTEMP_:
+ case OMP_CLAUSE__REDUCTEMP_:
do_in = true;
break;
@@ -5495,9 +6219,15 @@ lower_send_clauses (tree clauses, gimple
break;
case OMP_CLAUSE_REDUCTION:
+ case OMP_CLAUSE_IN_REDUCTION:
do_in = true;
if (val == OMP_CLAUSE_DECL (c))
- do_out = !(by_ref || omp_is_reference (val));
+ {
+ if (is_task_ctx (ctx))
+ by_ref = use_pointer_for_field (val, ctx);
+ else
+ do_out = !(by_ref || omp_is_reference (val));
+ }
else
by_ref = TREE_CODE (TREE_TYPE (val)) == ARRAY_TYPE;
break;
@@ -5802,30 +6532,55 @@ maybe_catch_exception (gimple_seq body)
cancellation in the implicit barrier. */
static void
-maybe_add_implicit_barrier_cancel (omp_context *ctx, gimple_seq *body)
+maybe_add_implicit_barrier_cancel (omp_context *ctx, gimple *omp_return,
+ gimple_seq *body)
{
- gimple *omp_return = gimple_seq_last_stmt (*body);
gcc_assert (gimple_code (omp_return) == GIMPLE_OMP_RETURN);
if (gimple_omp_return_nowait_p (omp_return))
return;
- if (ctx->outer
- && gimple_code (ctx->outer->stmt) == GIMPLE_OMP_PARALLEL
- && ctx->outer->cancellable)
- {
- tree fndecl = builtin_decl_explicit (BUILT_IN_GOMP_CANCEL);
- tree c_bool_type = TREE_TYPE (TREE_TYPE (fndecl));
- tree lhs = create_tmp_var (c_bool_type);
- gimple_omp_return_set_lhs (omp_return, lhs);
- tree fallthru_label = create_artificial_label (UNKNOWN_LOCATION);
- gimple *g = gimple_build_cond (NE_EXPR, lhs,
- fold_convert (c_bool_type,
- boolean_false_node),
- ctx->outer->cancel_label, fallthru_label);
- gimple_seq_add_stmt (body, g);
- gimple_seq_add_stmt (body, gimple_build_label (fallthru_label));
+ for (omp_context *outer = ctx->outer; outer; outer = outer->outer)
+ if (gimple_code (outer->stmt) == GIMPLE_OMP_PARALLEL
+ && outer->cancellable)
+ {
+ tree fndecl = builtin_decl_explicit (BUILT_IN_GOMP_CANCEL);
+ tree c_bool_type = TREE_TYPE (TREE_TYPE (fndecl));
+ tree lhs = create_tmp_var (c_bool_type);
+ gimple_omp_return_set_lhs (omp_return, lhs);
+ tree fallthru_label = create_artificial_label (UNKNOWN_LOCATION);
+ gimple *g = gimple_build_cond (NE_EXPR, lhs,
+ fold_convert (c_bool_type,
+ boolean_false_node),
+ outer->cancel_label, fallthru_label);
+ gimple_seq_add_stmt (body, g);
+ gimple_seq_add_stmt (body, gimple_build_label (fallthru_label));
+ }
+ else if (gimple_code (outer->stmt) != GIMPLE_OMP_TASKGROUP)
+ return;
+}
+
+/* Find the first task_reduction or reduction clause or return NULL
+ if there are none. */
+
+static inline tree
+omp_task_reductions_find_first (tree clauses, enum tree_code code,
+ enum omp_clause_code ccode)
+{
+ while (1)
+ {
+ clauses = omp_find_clause (clauses, ccode);
+ if (clauses == NULL_TREE)
+ return NULL_TREE;
+ if (ccode != OMP_CLAUSE_REDUCTION
+ || code == OMP_TASKLOOP
+ || OMP_CLAUSE_REDUCTION_TASK (clauses))
+ return clauses;
+ clauses = OMP_CLAUSE_CHAIN (clauses);
}
}
+static void lower_omp_task_reductions (omp_context *, enum tree_code, tree,
+ gimple_seq *, gimple_seq *);
+
/* Lower the OpenMP sections directive in the current statement in GSI_P.
CTX is the enclosing OMP context for the current statement. */
@@ -5837,7 +6592,7 @@ lower_omp_sections (gimple_stmt_iterator
gomp_sections *stmt;
gimple *t;
gbind *new_stmt, *bind;
- gimple_seq ilist, dlist, olist, new_body;
+ gimple_seq ilist, dlist, olist, tred_dlist = NULL, new_body;
stmt = as_a <gomp_sections *> (gsi_stmt (*gsi_p));
@@ -5845,6 +6600,27 @@ lower_omp_sections (gimple_stmt_iterator
dlist = NULL;
ilist = NULL;
+
+ tree rclauses
+ = omp_task_reductions_find_first (gimple_omp_sections_clauses (stmt),
+ OMP_SECTIONS, OMP_CLAUSE_REDUCTION);
+ tree rtmp = NULL_TREE;
+ if (rclauses)
+ {
+ tree type = build_pointer_type (pointer_sized_int_node);
+ tree temp = create_tmp_var (type);
+ tree c = build_omp_clause (UNKNOWN_LOCATION, OMP_CLAUSE__REDUCTEMP_);
+ OMP_CLAUSE_DECL (c) = temp;
+ OMP_CLAUSE_CHAIN (c) = gimple_omp_sections_clauses (stmt);
+ gimple_omp_sections_set_clauses (stmt, c);
+ lower_omp_task_reductions (ctx, OMP_SECTIONS,
+ gimple_omp_sections_clauses (stmt),
+ &ilist, &tred_dlist);
+ rclauses = c;
+ rtmp = make_ssa_name (type);
+ gimple_seq_add_stmt (&ilist, gimple_build_assign (rtmp, temp));
+ }
+
lower_rec_input_clauses (gimple_omp_sections_clauses (stmt),
&ilist, &dlist, ctx, NULL);
@@ -5916,7 +6692,11 @@ lower_omp_sections (gimple_stmt_iterator
OMP_CLAUSE_NOWAIT) != NULL_TREE;
t = gimple_build_omp_return (nowait);
gimple_seq_add_stmt (&new_body, t);
- maybe_add_implicit_barrier_cancel (ctx, &new_body);
+ gimple_seq_add_seq (&new_body, tred_dlist);
+ maybe_add_implicit_barrier_cancel (ctx, t, &new_body);
+
+ if (rclauses)
+ OMP_CLAUSE_DECL (rclauses) = rtmp;
gimple_bind_set_body (new_stmt, new_body);
}
@@ -6078,7 +6858,7 @@ lower_omp_single (gimple_stmt_iterator *
OMP_CLAUSE_NOWAIT) != NULL_TREE;
gimple *g = gimple_build_omp_return (nowait);
gimple_seq_add_stmt (&bind_body_tail, g);
- maybe_add_implicit_barrier_cancel (ctx, &bind_body_tail);
+ maybe_add_implicit_barrier_cancel (ctx, g, &bind_body_tail);
if (ctx->record_type)
{
gimple_stmt_iterator gsi = gsi_start (bind_body_tail);
@@ -6140,6 +6920,604 @@ lower_omp_master (gimple_stmt_iterator *
BLOCK_VARS (block) = ctx->block_vars;
}
+/* Helper function for lower_omp_task_reductions. For a specific PASS
+ find out the current clause it should be processed, or return false
+ if all have been processed already. */
+
+static inline bool
+omp_task_reduction_iterate (int pass, enum tree_code code,
+ enum omp_clause_code ccode, tree *c, tree *decl,
+ tree *type, tree *next)
+{
+ for (; *c; *c = omp_find_clause (OMP_CLAUSE_CHAIN (*c), ccode))
+ {
+ if (ccode == OMP_CLAUSE_REDUCTION
+ && code != OMP_TASKLOOP
+ && !OMP_CLAUSE_REDUCTION_TASK (*c))
+ continue;
+ *decl = OMP_CLAUSE_DECL (*c);
+ *type = TREE_TYPE (*decl);
+ if (TREE_CODE (*decl) == MEM_REF)
+ {
+ if (pass != 1)
+ continue;
+ }
+ else
+ {
+ if (omp_is_reference (*decl))
+ *type = TREE_TYPE (*type);
+ if (pass != (!TREE_CONSTANT (TYPE_SIZE_UNIT (*type))))
+ continue;
+ }
+ *next = omp_find_clause (OMP_CLAUSE_CHAIN (*c), ccode);
+ return true;
+ }
+ *decl = NULL_TREE;
+ *type = NULL_TREE;
+ *next = NULL_TREE;
+ return false;
+}
+
+/* Lower task_reduction and reduction clauses (the latter unless CODE is
+ OMP_TASKGROUP only with task modifier). Register mapping of those in
+ START sequence and reducing them and unregister them in the END sequence. */
+
+static void
+lower_omp_task_reductions (omp_context *ctx, enum tree_code code, tree clauses,
+ gimple_seq *start, gimple_seq *end)
+{
+ enum omp_clause_code ccode
+ = (code == OMP_TASKGROUP
+ ? OMP_CLAUSE_TASK_REDUCTION : OMP_CLAUSE_REDUCTION);
+ tree cancellable = NULL_TREE;
+ clauses = omp_task_reductions_find_first (clauses, code, ccode);
+ if (clauses == NULL_TREE)
+ return;
+ if (code == OMP_FOR || code == OMP_SECTIONS)
+ {
+ for (omp_context *outer = ctx->outer; outer; outer = outer->outer)
+ if (gimple_code (outer->stmt) == GIMPLE_OMP_PARALLEL
+ && outer->cancellable)
+ {
+ cancellable = error_mark_node;
+ break;
+ }
+ else if (gimple_code (outer->stmt) != GIMPLE_OMP_TASKGROUP)
+ break;
+ }
+ tree record_type = lang_hooks.types.make_type (RECORD_TYPE);
+ tree *last = &TYPE_FIELDS (record_type);
+ unsigned cnt = 0;
+ if (cancellable)
+ {
+ tree field = build_decl (UNKNOWN_LOCATION, FIELD_DECL, NULL_TREE,
+ ptr_type_node);
+ tree ifield = build_decl (UNKNOWN_LOCATION, FIELD_DECL, NULL_TREE,
+ integer_type_node);
+ *last = field;
+ DECL_CHAIN (field) = ifield;
+ last = &DECL_CHAIN (ifield);
+ }
+ for (int pass = 0; pass < 2; pass++)
+ {
+ tree decl, type, next;
+ for (tree c = clauses;
+ omp_task_reduction_iterate (pass, code, ccode,
+ &c, &decl, &type, &next); c = next)
+ {
+ ++cnt;
+ tree new_type = type;
+ if (ctx->outer)
+ new_type = remap_type (type, &ctx->outer->cb);
+ tree field
+ = build_decl (OMP_CLAUSE_LOCATION (c), FIELD_DECL,
+ DECL_P (decl) ? DECL_NAME (decl) : NULL_TREE,
+ new_type);
+ if (DECL_P (decl) && type == TREE_TYPE (decl))
+ {
+ SET_DECL_ALIGN (field, DECL_ALIGN (decl));
+ DECL_USER_ALIGN (field) = DECL_USER_ALIGN (decl);
+ TREE_THIS_VOLATILE (field) = TREE_THIS_VOLATILE (decl);
+ }
+ else
+ SET_DECL_ALIGN (field, TYPE_ALIGN (type));
+ DECL_CONTEXT (field) = record_type;
+ *last = field;
+ last = &DECL_CHAIN (field);
+ tree bfield
+ = build_decl (OMP_CLAUSE_LOCATION (c), FIELD_DECL, NULL_TREE,
+ boolean_type_node);
+ DECL_CONTEXT (bfield) = record_type;
+ *last = bfield;
+ last = &DECL_CHAIN (bfield);
+ }
+ }
+ *last = NULL_TREE;
+ layout_type (record_type);
+
+ /* Build up an array which registers with the runtime all the reductions
+ and deregisters them at the end. Format documented in libgomp/task.c. */
+ tree atype = build_array_type_nelts (pointer_sized_int_node, 7 + cnt * 3);
+ tree avar = create_tmp_var_raw (atype);
+ gimple_add_tmp_var (avar);
+ TREE_ADDRESSABLE (avar) = 1;
+ tree r = build4 (ARRAY_REF, pointer_sized_int_node, avar, size_zero_node,
+ NULL_TREE, NULL_TREE);
+ tree t = build_int_cst (pointer_sized_int_node, cnt);
+ gimple_seq_add_stmt (start, gimple_build_assign (r, t));
+ gimple_seq seq = NULL;
+ tree sz = fold_convert (pointer_sized_int_node,
+ TYPE_SIZE_UNIT (record_type));
+ int cachesz = 64;
+ sz = fold_build2 (PLUS_EXPR, pointer_sized_int_node, sz,
+ build_int_cst (pointer_sized_int_node, cachesz - 1));
+ sz = fold_build2 (BIT_AND_EXPR, pointer_sized_int_node, sz,
+ build_int_cst (pointer_sized_int_node, ~(cachesz - 1)));
+ ctx->task_reductions.create (1 + cnt);
+ ctx->task_reduction_map = new hash_map<tree, unsigned>;
+ ctx->task_reductions.quick_push (TREE_CODE (sz) == INTEGER_CST
+ ? sz : NULL_TREE);
+ sz = force_gimple_operand (sz, &seq, true, NULL_TREE);
+ gimple_seq_add_seq (start, seq);
+ r = build4 (ARRAY_REF, pointer_sized_int_node, avar, size_one_node,
+ NULL_TREE, NULL_TREE);
+ gimple_seq_add_stmt (start, gimple_build_assign (r, sz));
+ r = build4 (ARRAY_REF, pointer_sized_int_node, avar, size_int (2),
+ NULL_TREE, NULL_TREE);
+ t = build_int_cst (pointer_sized_int_node,
+ MAX (TYPE_ALIGN_UNIT (record_type), (unsigned) cachesz));
+ gimple_seq_add_stmt (start, gimple_build_assign (r, t));
+ r = build4 (ARRAY_REF, pointer_sized_int_node, avar, size_int (3),
+ NULL_TREE, NULL_TREE);
+ t = build_int_cst (pointer_sized_int_node, -1);
+ gimple_seq_add_stmt (start, gimple_build_assign (r, t));
+ r = build4 (ARRAY_REF, pointer_sized_int_node, avar, size_int (4),
+ NULL_TREE, NULL_TREE);
+ t = build_int_cst (pointer_sized_int_node, 0);
+ gimple_seq_add_stmt (start, gimple_build_assign (r, t));
+
+ /* In end, build a loop that iterates from 0 to < omp_get_num_threads ()
+ and for each task reduction checks a bool right after the private variable
+ within that thread's chunk; if the bool is clear, it hasn't been
+ initialized and thus isn't going to be reduced nor destructed, otherwise
+ reduce and destruct it. */
+ tree idx = create_tmp_var (size_type_node);
+ gimple_seq_add_stmt (end, gimple_build_assign (idx, size_zero_node));
+ tree num_thr_sz = create_tmp_var (size_type_node);
+ tree lab1 = create_artificial_label (UNKNOWN_LOCATION);
+ tree lab2 = create_artificial_label (UNKNOWN_LOCATION);
+ tree lab3 = NULL_TREE;
+ gimple *g;
+ if (code == OMP_FOR || code == OMP_SECTIONS)
+ {
+ /* For worksharing constructs, only perform it in the master thread,
+ with the exception of cancelled implicit barriers - then only handle
+ the current thread. */
+ tree lab4 = create_artificial_label (UNKNOWN_LOCATION);
+ t = builtin_decl_explicit (BUILT_IN_OMP_GET_THREAD_NUM);
+ tree thr_num = create_tmp_var (integer_type_node);
+ g = gimple_build_call (t, 0);
+ gimple_call_set_lhs (g, thr_num);
+ gimple_seq_add_stmt (end, g);
+ if (cancellable)
+ {
+ tree c;
+ tree lab5 = create_artificial_label (UNKNOWN_LOCATION);
+ tree lab6 = create_artificial_label (UNKNOWN_LOCATION);
+ lab3 = create_artificial_label (UNKNOWN_LOCATION);
+ if (code == OMP_FOR)
+ c = gimple_omp_for_clauses (ctx->stmt);
+ else if (code == OMP_SECTIONS)
+ c = gimple_omp_sections_clauses (ctx->stmt);
+ c = OMP_CLAUSE_DECL (omp_find_clause (c, OMP_CLAUSE__REDUCTEMP_));
+ cancellable = c;
+ g = gimple_build_cond (NE_EXPR, c, build_zero_cst (TREE_TYPE (c)),
+ lab5, lab6);
+ gimple_seq_add_stmt (end, g);
+ gimple_seq_add_stmt (end, gimple_build_label (lab5));
+ g = gimple_build_assign (idx, NOP_EXPR, thr_num);
+ gimple_seq_add_stmt (end, g);
+ g = gimple_build_assign (num_thr_sz, PLUS_EXPR, idx,
+ build_one_cst (TREE_TYPE (idx)));
+ gimple_seq_add_stmt (end, g);
+ gimple_seq_add_stmt (end, gimple_build_goto (lab3));
+ gimple_seq_add_stmt (end, gimple_build_label (lab6));
+ }
+ g = gimple_build_cond (NE_EXPR, thr_num, integer_zero_node, lab2, lab4);
+ gimple_seq_add_stmt (end, g);
+ gimple_seq_add_stmt (end, gimple_build_label (lab4));
+ }
+ if (code != OMP_PARALLEL)
+ {
+ t = builtin_decl_explicit (BUILT_IN_OMP_GET_NUM_THREADS);
+ tree num_thr = create_tmp_var (integer_type_node);
+ g = gimple_build_call (t, 0);
+ gimple_call_set_lhs (g, num_thr);
+ gimple_seq_add_stmt (end, g);
+ g = gimple_build_assign (num_thr_sz, NOP_EXPR, num_thr);
+ gimple_seq_add_stmt (end, g);
+ if (cancellable)
+ gimple_seq_add_stmt (end, gimple_build_label (lab3));
+ }
+ else
+ {
+ tree c = omp_find_clause (gimple_omp_parallel_clauses (ctx->stmt),
+ OMP_CLAUSE__REDUCTEMP_);
+ t = fold_convert (pointer_sized_int_node, OMP_CLAUSE_DECL (c));
+ t = fold_convert (size_type_node, t);
+ gimplify_assign (num_thr_sz, t, end);
+ }
+ t = build4 (ARRAY_REF, pointer_sized_int_node, avar, size_int (2),
+ NULL_TREE, NULL_TREE);
+ tree data = create_tmp_var (pointer_sized_int_node);
+ gimple_seq_add_stmt (end, gimple_build_assign (data, t));
+ gimple_seq_add_stmt (end, gimple_build_label (lab1));
+ tree ptr;
+ if (TREE_CODE (TYPE_SIZE_UNIT (record_type)) == INTEGER_CST)
+ ptr = create_tmp_var (build_pointer_type (record_type));
+ else
+ ptr = create_tmp_var (ptr_type_node);
+ gimple_seq_add_stmt (end, gimple_build_assign (ptr, NOP_EXPR, data));
+
+ tree field = TYPE_FIELDS (record_type);
+ cnt = 0;
+ if (cancellable)
+ field = DECL_CHAIN (DECL_CHAIN (field));
+ for (int pass = 0; pass < 2; pass++)
+ {
+ tree decl, type, next;
+ for (tree c = clauses;
+ omp_task_reduction_iterate (pass, code, ccode,
+ &c, &decl, &type, &next); c = next)
+ {
+ tree var = decl, ref;
+ if (TREE_CODE (decl) == MEM_REF)
+ {
+ var = TREE_OPERAND (var, 0);
+ if (TREE_CODE (var) == POINTER_PLUS_EXPR)
+ var = TREE_OPERAND (var, 0);
+ tree v = var;
+ if (TREE_CODE (var) == ADDR_EXPR)
+ var = TREE_OPERAND (var, 0);
+ else if (TREE_CODE (var) == INDIRECT_REF)
+ var = TREE_OPERAND (var, 0);
+ tree orig_var = var;
+ if (is_variable_sized (var))
+ {
+ gcc_assert (DECL_HAS_VALUE_EXPR_P (var));
+ var = DECL_VALUE_EXPR (var);
+ gcc_assert (TREE_CODE (var) == INDIRECT_REF);
+ var = TREE_OPERAND (var, 0);
+ gcc_assert (DECL_P (var));
+ }
+ t = ref = maybe_lookup_decl_in_outer_ctx (var, ctx);
+ if (orig_var != var)
+ gcc_assert (TREE_CODE (v) == ADDR_EXPR);
+ else if (TREE_CODE (v) == ADDR_EXPR)
+ t = build_fold_addr_expr (t);
+ else if (TREE_CODE (v) == INDIRECT_REF)
+ t = build_fold_indirect_ref (t);
+ if (TREE_CODE (TREE_OPERAND (decl, 0)) == POINTER_PLUS_EXPR)
+ {
+ tree b = TREE_OPERAND (TREE_OPERAND (decl, 0), 1);
+ b = maybe_lookup_decl_in_outer_ctx (b, ctx);
+ t = fold_build2 (POINTER_PLUS_EXPR, TREE_TYPE (t), t, b);
+ }
+ if (!integer_zerop (TREE_OPERAND (decl, 1)))
+ t = fold_build2 (POINTER_PLUS_EXPR, TREE_TYPE (t), t,
+ fold_convert (size_type_node,
+ TREE_OPERAND (decl, 1)));
+ }
+ else
+ {
+ t = ref = maybe_lookup_decl_in_outer_ctx (var, ctx);
+ if (!omp_is_reference (decl))
+ t = build_fold_addr_expr (t);
+ }
+ t = fold_convert (pointer_sized_int_node, t);
+ seq = NULL;
+ t = force_gimple_operand (t, &seq, true, NULL_TREE);
+ gimple_seq_add_seq (start, seq);
+ r = build4 (ARRAY_REF, pointer_sized_int_node, avar,
+ size_int (7 + cnt * 3), NULL_TREE, NULL_TREE);
+ gimple_seq_add_stmt (start, gimple_build_assign (r, t));
+ t = unshare_expr (byte_position (field));
+ t = fold_convert (pointer_sized_int_node, t);
+ ctx->task_reduction_map->put (c, cnt);
+ ctx->task_reductions.quick_push (TREE_CODE (t) == INTEGER_CST
+ ? t : NULL_TREE);
+ seq = NULL;
+ t = force_gimple_operand (t, &seq, true, NULL_TREE);
+ gimple_seq_add_seq (start, seq);
+ r = build4 (ARRAY_REF, pointer_sized_int_node, avar,
+ size_int (7 + cnt * 3 + 1), NULL_TREE, NULL_TREE);
+ gimple_seq_add_stmt (start, gimple_build_assign (r, t));
+
+ tree bfield = DECL_CHAIN (field);
+ tree cond;
+ if (code == OMP_PARALLEL || code == OMP_FOR || code == OMP_SECTIONS)
+ /* In parallel or worksharing all threads unconditionally
+ initialize all their task reduction private variables. */
+ cond = boolean_true_node;
+ else if (TREE_TYPE (ptr) == ptr_type_node)
+ {
+ cond = build2 (POINTER_PLUS_EXPR, ptr_type_node, ptr,
+ unshare_expr (byte_position (bfield)));
+ seq = NULL;
+ cond = force_gimple_operand (cond, &seq, true, NULL_TREE);
+ gimple_seq_add_seq (end, seq);
+ tree pbool = build_pointer_type (TREE_TYPE (bfield));
+ cond = build2 (MEM_REF, TREE_TYPE (bfield), cond,
+ build_int_cst (pbool, 0));
+ }
+ else
+ cond = build3 (COMPONENT_REF, TREE_TYPE (bfield),
+ build_simple_mem_ref (ptr), bfield, NULL_TREE);
+ tree lab3 = create_artificial_label (UNKNOWN_LOCATION);
+ tree lab4 = create_artificial_label (UNKNOWN_LOCATION);
+ tree condv = create_tmp_var (boolean_type_node);
+ gimple_seq_add_stmt (end, gimple_build_assign (condv, cond));
+ g = gimple_build_cond (NE_EXPR, condv, boolean_false_node,
+ lab3, lab4);
+ gimple_seq_add_stmt (end, g);
+ gimple_seq_add_stmt (end, gimple_build_label (lab3));
+ if (cancellable && OMP_CLAUSE_REDUCTION_PLACEHOLDER (c) == NULL_TREE)
+ {
+ /* If this reduction doesn't need destruction and parallel
+ has been cancelled, there is nothing to do for this
+ reduction, so jump around the merge operation. */
+ tree lab5 = create_artificial_label (UNKNOWN_LOCATION);
+ g = gimple_build_cond (NE_EXPR, cancellable,
+ build_zero_cst (TREE_TYPE (cancellable)),
+ lab4, lab5);
+ gimple_seq_add_stmt (end, g);
+ gimple_seq_add_stmt (end, gimple_build_label (lab5));
+ }
+
+ tree new_var;
+ if (TREE_TYPE (ptr) == ptr_type_node)
+ {
+ new_var = build2 (POINTER_PLUS_EXPR, ptr_type_node, ptr,
+ unshare_expr (byte_position (field)));
+ seq = NULL;
+ new_var = force_gimple_operand (new_var, &seq, true, NULL_TREE);
+ gimple_seq_add_seq (end, seq);
+ tree pbool = build_pointer_type (TREE_TYPE (field));
+ new_var = build2 (MEM_REF, TREE_TYPE (field), new_var,
+ build_int_cst (pbool, 0));
+ }
+ else
+ new_var = build3 (COMPONENT_REF, TREE_TYPE (field),
+ build_simple_mem_ref (ptr), field, NULL_TREE);
+
+ enum tree_code rcode = OMP_CLAUSE_REDUCTION_CODE (c);
+ if (TREE_CODE (decl) != MEM_REF && omp_is_reference (decl))
+ ref = build_simple_mem_ref (ref);
+ /* reduction(-:var) sums up the partial results, so it acts
+ identically to reduction(+:var). */
+ if (rcode == MINUS_EXPR)
+ rcode = PLUS_EXPR;
+ if (TREE_CODE (decl) == MEM_REF)
+ {
+ tree type = TREE_TYPE (new_var);
+ tree v = TYPE_MAX_VALUE (TYPE_DOMAIN (type));
+ tree i = create_tmp_var (TREE_TYPE (v));
+ tree ptype = build_pointer_type (TREE_TYPE (type));
+ if (DECL_P (v))
+ {
+ v = maybe_lookup_decl_in_outer_ctx (v, ctx);
+ tree vv = create_tmp_var (TREE_TYPE (v));
+ gimplify_assign (vv, v, start);
+ v = vv;
+ }
+ ref = build4 (ARRAY_REF, pointer_sized_int_node, avar,
+ size_int (7 + cnt * 3), NULL_TREE, NULL_TREE);
+ new_var = build_fold_addr_expr (new_var);
+ new_var = fold_convert (ptype, new_var);
+ ref = fold_convert (ptype, ref);
+ tree m = create_tmp_var (ptype);
+ gimplify_assign (m, new_var, end);
+ new_var = m;
+ m = create_tmp_var (ptype);
+ gimplify_assign (m, ref, end);
+ ref = m;
+ gimplify_assign (i, build_int_cst (TREE_TYPE (v), 0), end);
+ tree body = create_artificial_label (UNKNOWN_LOCATION);
+ tree endl = create_artificial_label (UNKNOWN_LOCATION);
+ gimple_seq_add_stmt (end, gimple_build_label (body));
+ tree priv = build_simple_mem_ref (new_var);
+ tree out = build_simple_mem_ref (ref);
+ if (OMP_CLAUSE_REDUCTION_PLACEHOLDER (c))
+ {
+ tree placeholder = OMP_CLAUSE_REDUCTION_PLACEHOLDER (c);
+ tree decl_placeholder
+ = OMP_CLAUSE_REDUCTION_DECL_PLACEHOLDER (c);
+ tree lab6 = NULL_TREE;
+ if (cancellable)
+ {
+ /* If this reduction needs destruction and parallel
+ has been cancelled, jump around the merge operation
+ to the destruction. */
+ tree lab5 = create_artificial_label (UNKNOWN_LOCATION);
+ lab6 = create_artificial_label (UNKNOWN_LOCATION);
+ tree zero = build_zero_cst (TREE_TYPE (cancellable));
+ g = gimple_build_cond (NE_EXPR, cancellable, zero,
+ lab6, lab5);
+ gimple_seq_add_stmt (end, g);
+ gimple_seq_add_stmt (end, gimple_build_label (lab5));
+ }
+ SET_DECL_VALUE_EXPR (placeholder, out);
+ DECL_HAS_VALUE_EXPR_P (placeholder) = 1;
+ SET_DECL_VALUE_EXPR (decl_placeholder, priv);
+ DECL_HAS_VALUE_EXPR_P (decl_placeholder) = 1;
+ lower_omp (&OMP_CLAUSE_REDUCTION_GIMPLE_MERGE (c), ctx);
+ gimple_seq_add_seq (end,
+ OMP_CLAUSE_REDUCTION_GIMPLE_MERGE (c));
+ OMP_CLAUSE_REDUCTION_GIMPLE_MERGE (c) = NULL;
+ if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_TASK_REDUCTION)
+ {
+ OMP_CLAUSE_REDUCTION_PLACEHOLDER (c) = NULL;
+ OMP_CLAUSE_REDUCTION_DECL_PLACEHOLDER (c) = NULL;
+ }
+ if (cancellable)
+ gimple_seq_add_stmt (end, gimple_build_label (lab6));
+ tree x = lang_hooks.decls.omp_clause_dtor (c, priv);
+ if (x)
+ {
+ gimple_seq tseq = NULL;
+ gimplify_stmt (&x, &tseq);
+ gimple_seq_add_seq (end, tseq);
+ }
+ }
+ else
+ {
+ tree x = build2 (rcode, TREE_TYPE (out), out, priv);
+ out = unshare_expr (out);
+ gimplify_assign (out, x, end);
+ }
+ gimple *g
+ = gimple_build_assign (new_var, POINTER_PLUS_EXPR, new_var,
+ TYPE_SIZE_UNIT (TREE_TYPE (type)));
+ gimple_seq_add_stmt (end, g);
+ g = gimple_build_assign (ref, POINTER_PLUS_EXPR, ref,
+ TYPE_SIZE_UNIT (TREE_TYPE (type)));
+ gimple_seq_add_stmt (end, g);
+ g = gimple_build_assign (i, PLUS_EXPR, i,
+ build_int_cst (TREE_TYPE (i), 1));
+ gimple_seq_add_stmt (end, g);
+ g = gimple_build_cond (LE_EXPR, i, v, body, endl);
+ gimple_seq_add_stmt (end, g);
+ gimple_seq_add_stmt (end, gimple_build_label (endl));
+ }
+ else if (OMP_CLAUSE_REDUCTION_PLACEHOLDER (c))
+ {
+ tree placeholder = OMP_CLAUSE_REDUCTION_PLACEHOLDER (c);
+ tree oldv = NULL_TREE;
+ tree lab6 = NULL_TREE;
+ if (cancellable)
+ {
+ /* If this reduction needs destruction and parallel
+ has been cancelled, jump around the merge operation
+ to the destruction. */
+ tree lab5 = create_artificial_label (UNKNOWN_LOCATION);
+ lab6 = create_artificial_label (UNKNOWN_LOCATION);
+ tree zero = build_zero_cst (TREE_TYPE (cancellable));
+ g = gimple_build_cond (NE_EXPR, cancellable, zero,
+ lab6, lab5);
+ gimple_seq_add_stmt (end, g);
+ gimple_seq_add_stmt (end, gimple_build_label (lab5));
+ }
+ if (omp_is_reference (decl)
+ && !useless_type_conversion_p (TREE_TYPE (placeholder),
+ TREE_TYPE (ref)))
+ ref = build_fold_addr_expr_loc (OMP_CLAUSE_LOCATION (c), ref);
+ ref = build_fold_addr_expr_loc (OMP_CLAUSE_LOCATION (c), ref);
+ tree refv = create_tmp_var (TREE_TYPE (ref));
+ gimplify_assign (refv, ref, end);
+ ref = build_simple_mem_ref_loc (OMP_CLAUSE_LOCATION (c), refv);
+ SET_DECL_VALUE_EXPR (placeholder, ref);
+ DECL_HAS_VALUE_EXPR_P (placeholder) = 1;
+ tree d = maybe_lookup_decl (decl, ctx);
+ gcc_assert (d);
+ if (DECL_HAS_VALUE_EXPR_P (d))
+ oldv = DECL_VALUE_EXPR (d);
+ if (omp_is_reference (var))
+ {
+ tree v = fold_convert (TREE_TYPE (d),
+ build_fold_addr_expr (new_var));
+ SET_DECL_VALUE_EXPR (d, v);
+ }
+ else
+ SET_DECL_VALUE_EXPR (d, new_var);
+ DECL_HAS_VALUE_EXPR_P (d) = 1;
+ lower_omp (&OMP_CLAUSE_REDUCTION_GIMPLE_MERGE (c), ctx);
+ if (oldv)
+ SET_DECL_VALUE_EXPR (d, oldv);
+ else
+ {
+ SET_DECL_VALUE_EXPR (d, NULL_TREE);
+ DECL_HAS_VALUE_EXPR_P (d) = 0;
+ }
+ gimple_seq_add_seq (end, OMP_CLAUSE_REDUCTION_GIMPLE_MERGE (c));
+ OMP_CLAUSE_REDUCTION_GIMPLE_MERGE (c) = NULL;
+ if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_TASK_REDUCTION)
+ OMP_CLAUSE_REDUCTION_PLACEHOLDER (c) = NULL;
+ if (cancellable)
+ gimple_seq_add_stmt (end, gimple_build_label (lab6));
+ tree x = lang_hooks.decls.omp_clause_dtor (c, new_var);
+ if (x)
+ {
+ gimple_seq tseq = NULL;
+ gimplify_stmt (&x, &tseq);
+ gimple_seq_add_seq (end, tseq);
+ }
+ }
+ else
+ {
+ tree x = build2 (rcode, TREE_TYPE (ref), ref, new_var);
+ ref = unshare_expr (ref);
+ gimplify_assign (ref, x, end);
+ }
+ gimple_seq_add_stmt (end, gimple_build_label (lab4));
+ ++cnt;
+ field = DECL_CHAIN (bfield);
+ }
+ }
+
+ if (code == OMP_TASKGROUP)
+ {
+ t = builtin_decl_explicit (BUILT_IN_GOMP_TASKGROUP_REDUCTION_REGISTER);
+ g = gimple_build_call (t, 1, build_fold_addr_expr (avar));
+ gimple_seq_add_stmt (start, g);
+ }
+ else
+ {
+ tree c;
+ if (code == OMP_FOR)
+ c = gimple_omp_for_clauses (ctx->stmt);
+ else if (code == OMP_SECTIONS)
+ c = gimple_omp_sections_clauses (ctx->stmt);
+ else
+ c = gimple_omp_taskreg_clauses (ctx->stmt);
+ c = omp_find_clause (c, OMP_CLAUSE__REDUCTEMP_);
+ t = fold_convert (TREE_TYPE (OMP_CLAUSE_DECL (c)),
+ build_fold_addr_expr (avar));
+ gimplify_assign (OMP_CLAUSE_DECL (c), t, start);
+ }
+
+ gimple_seq_add_stmt (end, gimple_build_assign (data, PLUS_EXPR, data, sz));
+ gimple_seq_add_stmt (end, gimple_build_assign (idx, PLUS_EXPR, idx,
+ size_one_node));
+ g = gimple_build_cond (NE_EXPR, idx, num_thr_sz, lab1, lab2);
+ gimple_seq_add_stmt (end, g);
+ gimple_seq_add_stmt (end, gimple_build_label (lab2));
+ if (code == OMP_FOR || code == OMP_SECTIONS)
+ {
+ enum built_in_function bfn
+ = BUILT_IN_GOMP_WORKSHARE_TASK_REDUCTION_UNREGISTER;
+ t = builtin_decl_explicit (bfn);
+ tree c_bool_type = TREE_VALUE (TYPE_ARG_TYPES (TREE_TYPE (t)));
+ tree arg;
+ if (cancellable)
+ {
+ arg = create_tmp_var (c_bool_type);
+ gimple_seq_add_stmt (end, gimple_build_assign (arg, NOP_EXPR,
+ cancellable));
+ }
+ else
+ arg = build_int_cst (c_bool_type, 0);
+ g = gimple_build_call (t, 1, arg);
+ }
+ else
+ {
+ t = builtin_decl_explicit (BUILT_IN_GOMP_TASKGROUP_REDUCTION_UNREGISTER);
+ g = gimple_build_call (t, 1, build_fold_addr_expr (avar));
+ }
+ gimple_seq_add_stmt (end, g);
+ t = build_constructor (atype, NULL);
+ TREE_THIS_VOLATILE (t) = 1;
+ gimple_seq_add_stmt (end, gimple_build_assign (avar, t));
+}
/* Expand code for an OpenMP taskgroup directive. */
@@ -6149,21 +7527,31 @@ lower_omp_taskgroup (gimple_stmt_iterato
gimple *stmt = gsi_stmt (*gsi_p);
gcall *x;
gbind *bind;
+ gimple_seq dseq = NULL;
tree block = make_node (BLOCK);
bind = gimple_build_bind (NULL, NULL, block);
gsi_replace (gsi_p, bind, true);
gimple_bind_add_stmt (bind, stmt);
+ push_gimplify_context ();
+
x = gimple_build_call (builtin_decl_explicit (BUILT_IN_GOMP_TASKGROUP_START),
0);
gimple_bind_add_stmt (bind, x);
+ lower_omp_task_reductions (ctx, OMP_TASKGROUP,
+ gimple_omp_taskgroup_clauses (stmt),
+ gimple_bind_body_ptr (bind), &dseq);
+
lower_omp (gimple_omp_body_ptr (stmt), ctx);
gimple_bind_add_seq (bind, gimple_omp_body (stmt));
gimple_omp_set_body (stmt, NULL);
gimple_bind_add_stmt (bind, gimple_build_omp_return (true));
+ gimple_bind_add_seq (bind, dseq);
+
+ pop_gimplify_context (bind);
gimple_bind_append_vars (bind, ctx->block_vars);
BLOCK_VARS (block) = ctx->block_vars;
@@ -6752,7 +8140,8 @@ lower_omp_for (gimple_stmt_iterator *gsi
struct omp_for_data fd, *fdp = NULL;
gomp_for *stmt = as_a <gomp_for *> (gsi_stmt (*gsi_p));
gbind *new_stmt;
- gimple_seq omp_for_body, body, dlist;
+ gimple_seq omp_for_body, body, dlist, tred_ilist = NULL, tred_dlist = NULL;
+ gimple_seq cnt_list = NULL;
gimple_seq oacc_head = NULL, oacc_tail = NULL;
size_t i;
@@ -6845,9 +8234,30 @@ lower_omp_for (gimple_stmt_iterator *gsi
/* The pre-body and input clauses go before the lowered GIMPLE_OMP_FOR. */
dlist = NULL;
body = NULL;
+ tree rclauses
+ = omp_task_reductions_find_first (gimple_omp_for_clauses (stmt), OMP_FOR,
+ OMP_CLAUSE_REDUCTION);
+ tree rtmp = NULL_TREE;
+ if (rclauses)
+ {
+ tree type = build_pointer_type (pointer_sized_int_node);
+ tree temp = create_tmp_var (type);
+ tree c = build_omp_clause (UNKNOWN_LOCATION, OMP_CLAUSE__REDUCTEMP_);
+ OMP_CLAUSE_DECL (c) = temp;
+ OMP_CLAUSE_CHAIN (c) = gimple_omp_for_clauses (stmt);
+ gimple_omp_for_set_clauses (stmt, c);
+ lower_omp_task_reductions (ctx, OMP_FOR,
+ gimple_omp_for_clauses (stmt),
+ &tred_ilist, &tred_dlist);
+ rclauses = c;
+ rtmp = make_ssa_name (type);
+ gimple_seq_add_stmt (&body, gimple_build_assign (rtmp, temp));
+ }
+
lower_rec_input_clauses (gimple_omp_for_clauses (stmt), &body, &dlist, ctx,
fdp);
- gimple_seq_add_seq (&body, gimple_omp_for_pre_body (stmt));
+ gimple_seq_add_seq (rclauses ? &tred_ilist : &body,
+ gimple_omp_for_pre_body (stmt));
lower_omp (gimple_omp_body_ptr (stmt), ctx);
@@ -6862,20 +8272,24 @@ lower_omp_for (gimple_stmt_iterator *gsi
{
rhs_p = gimple_omp_for_initial_ptr (stmt, i);
if (!is_gimple_min_invariant (*rhs_p))
- *rhs_p = get_formal_tmp_var (*rhs_p, &body);
+ *rhs_p = get_formal_tmp_var (*rhs_p, &cnt_list);
else if (TREE_CODE (*rhs_p) == ADDR_EXPR)
recompute_tree_invariant_for_addr_expr (*rhs_p);
rhs_p = gimple_omp_for_final_ptr (stmt, i);
if (!is_gimple_min_invariant (*rhs_p))
- *rhs_p = get_formal_tmp_var (*rhs_p, &body);
+ *rhs_p = get_formal_tmp_var (*rhs_p, &cnt_list);
else if (TREE_CODE (*rhs_p) == ADDR_EXPR)
recompute_tree_invariant_for_addr_expr (*rhs_p);
rhs_p = &TREE_OPERAND (gimple_omp_for_incr (stmt, i), 1);
if (!is_gimple_min_invariant (*rhs_p))
- *rhs_p = get_formal_tmp_var (*rhs_p, &body);
+ *rhs_p = get_formal_tmp_var (*rhs_p, &cnt_list);
}
+ if (rclauses)
+ gimple_seq_add_seq (&tred_ilist, cnt_list);
+ else
+ gimple_seq_add_seq (&body, cnt_list);
/* Once lowered, extract the bounds and clauses. */
omp_extract_for_data (stmt, &fd, NULL);
@@ -6922,13 +8336,26 @@ lower_omp_for (gimple_stmt_iterator *gsi
gimple_seq_add_seq (&body, dlist);
+ if (rclauses)
+ {
+ gimple_seq_add_seq (&tred_ilist, body);
+ body = tred_ilist;
+ }
+
body = maybe_catch_exception (body);
if (!phony_loop)
{
/* Region exit marker goes at the end of the loop body. */
- gimple_seq_add_stmt (&body, gimple_build_omp_return (fd.have_nowait));
- maybe_add_implicit_barrier_cancel (ctx, &body);
+ gimple *g = gimple_build_omp_return (fd.have_nowait);
+ gimple_seq_add_stmt (&body, g);
+
+ gimple_seq_add_seq (&body, tred_dlist);
+
+ maybe_add_implicit_barrier_cancel (ctx, g, &body);
+
+ if (rclauses)
+ OMP_CLAUSE_DECL (rclauses) = rtmp;
}
/* Add OpenACC joining and reduction markers just after the loop. */
@@ -7153,6 +8580,40 @@ create_task_copyfn (gomp_task *task_stmt
t = build2 (MODIFY_EXPR, TREE_TYPE (dst), dst, src);
append_to_statement_list (t, &list);
break;
+ case OMP_CLAUSE_REDUCTION:
+ case OMP_CLAUSE_IN_REDUCTION:
+ decl = OMP_CLAUSE_DECL (c);
+ if (TREE_CODE (decl) == MEM_REF)
+ {
+ decl = TREE_OPERAND (decl, 0);
+ if (TREE_CODE (decl) == POINTER_PLUS_EXPR)
+ decl = TREE_OPERAND (decl, 0);
+ if (TREE_CODE (decl) == INDIRECT_REF
+ || TREE_CODE (decl) == ADDR_EXPR)
+ decl = TREE_OPERAND (decl, 0);
+ }
+ key = (splay_tree_key) decl;
+ n = splay_tree_lookup (ctx->field_map, key);
+ if (n == NULL)
+ break;
+ f = (tree) n->value;
+ if (tcctx.cb.decl_map)
+ f = *tcctx.cb.decl_map->get (f);
+ n = splay_tree_lookup (ctx->sfield_map, key);
+ sf = (tree) n->value;
+ if (tcctx.cb.decl_map)
+ sf = *tcctx.cb.decl_map->get (sf);
+ src = build_simple_mem_ref_loc (loc, sarg);
+ src = omp_build_component_ref (src, sf);
+ if (decl != OMP_CLAUSE_DECL (c)
+ && TREE_CODE (TREE_TYPE (decl)) == REFERENCE_TYPE
+ && TREE_CODE (TREE_TYPE (TREE_TYPE (decl))) == POINTER_TYPE)
+ src = build_simple_mem_ref_loc (loc, src);
+ dst = build_simple_mem_ref_loc (loc, arg);
+ dst = omp_build_component_ref (dst, f);
+ t = build2 (MODIFY_EXPR, TREE_TYPE (dst), dst, src);
+ append_to_statement_list (t, &list);
+ break;
case OMP_CLAUSE__LOOPTEMP_:
/* Fields for first two _looptemp_ clauses are initialized by
GOMP_taskloop*, the rest are handled like firstprivate. */
@@ -7162,6 +8623,7 @@ create_task_copyfn (gomp_task *task_stmt
break;
}
/* FALLTHRU */
+ case OMP_CLAUSE__REDUCTEMP_:
case OMP_CLAUSE_FIRSTPRIVATE:
decl = OMP_CLAUSE_DECL (c);
if (is_variable_sized (decl))
@@ -7187,7 +8649,7 @@ create_task_copyfn (gomp_task *task_stmt
src = decl;
dst = build_simple_mem_ref_loc (loc, arg);
dst = omp_build_component_ref (dst, f);
- if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE__LOOPTEMP_)
+ if (OMP_CLAUSE_CODE (c) != OMP_CLAUSE_FIRSTPRIVATE)
t = build2 (MODIFY_EXPR, TREE_TYPE (dst), dst, src);
else
t = lang_hooks.decls.omp_clause_copy_ctor (c, dst, src);
@@ -7279,7 +8741,7 @@ lower_depend_clauses (tree *pclauses, gi
{
tree c, clauses;
gimple *g;
- size_t n_in = 0, n_out = 0, idx = 2, i;
+ size_t cnt[4] = { 0, 0, 0, 0 }, idx = 2, i;
clauses = omp_find_clause (*pclauses, OMP_CLAUSE_DEPEND);
gcc_assert (clauses);
@@ -7287,12 +8749,21 @@ lower_depend_clauses (tree *pclauses, gi
if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_DEPEND)
switch (OMP_CLAUSE_DEPEND_KIND (c))
{
+ case OMP_CLAUSE_DEPEND_LAST:
+ /* Lowering already done at gimplification. */
+ return;
case OMP_CLAUSE_DEPEND_IN:
- n_in++;
+ cnt[2]++;
break;
case OMP_CLAUSE_DEPEND_OUT:
case OMP_CLAUSE_DEPEND_INOUT:
- n_out++;
+ cnt[0]++;
+ break;
+ case OMP_CLAUSE_DEPEND_MUTEXINOUTSET:
+ cnt[1]++;
+ break;
+ case OMP_CLAUSE_DEPEND_DEPOBJ:
+ cnt[3]++;
break;
case OMP_CLAUSE_DEPEND_SOURCE:
case OMP_CLAUSE_DEPEND_SINK:
@@ -7300,25 +8771,61 @@ lower_depend_clauses (tree *pclauses, gi
default:
gcc_unreachable ();
}
- tree type = build_array_type_nelts (ptr_type_node, n_in + n_out + 2);
+ if (cnt[1] || cnt[3])
+ idx = 5;
+ size_t total = cnt[0] + cnt[1] + cnt[2] + cnt[3];
+ tree type = build_array_type_nelts (ptr_type_node, total + idx);
tree array = create_tmp_var (type);
TREE_ADDRESSABLE (array) = 1;
tree r = build4 (ARRAY_REF, ptr_type_node, array, size_int (0), NULL_TREE,
NULL_TREE);
- g = gimple_build_assign (r, build_int_cst (ptr_type_node, n_in + n_out));
- gimple_seq_add_stmt (iseq, g);
- r = build4 (ARRAY_REF, ptr_type_node, array, size_int (1), NULL_TREE,
- NULL_TREE);
- g = gimple_build_assign (r, build_int_cst (ptr_type_node, n_out));
+ if (idx == 5)
+ {
+ g = gimple_build_assign (r, build_int_cst (ptr_type_node, 0));
+ gimple_seq_add_stmt (iseq, g);
+ r = build4 (ARRAY_REF, ptr_type_node, array, size_int (1), NULL_TREE,
+ NULL_TREE);
+ }
+ g = gimple_build_assign (r, build_int_cst (ptr_type_node, total));
gimple_seq_add_stmt (iseq, g);
- for (i = 0; i < 2; i++)
+ for (i = 0; i < (idx == 5 ? 3 : 1); i++)
+ {
+ r = build4 (ARRAY_REF, ptr_type_node, array,
+ size_int (i + 1 + (idx == 5)), NULL_TREE, NULL_TREE);
+ g = gimple_build_assign (r, build_int_cst (ptr_type_node, cnt[i]));
+ gimple_seq_add_stmt (iseq, g);
+ }
+ for (i = 0; i < 4; i++)
{
- if ((i ? n_in : n_out) == 0)
+ if (cnt[i] == 0)
continue;
for (c = clauses; c; c = OMP_CLAUSE_CHAIN (c))
- if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_DEPEND
- && ((OMP_CLAUSE_DEPEND_KIND (c) != OMP_CLAUSE_DEPEND_IN) ^ i))
+ if (OMP_CLAUSE_CODE (c) != OMP_CLAUSE_DEPEND)
+ continue;
+ else
{
+ switch (OMP_CLAUSE_DEPEND_KIND (c))
+ {
+ case OMP_CLAUSE_DEPEND_IN:
+ if (i != 2)
+ continue;
+ break;
+ case OMP_CLAUSE_DEPEND_OUT:
+ case OMP_CLAUSE_DEPEND_INOUT:
+ if (i != 0)
+ continue;
+ break;
+ case OMP_CLAUSE_DEPEND_MUTEXINOUTSET:
+ if (i != 1)
+ continue;
+ break;
+ case OMP_CLAUSE_DEPEND_DEPOBJ:
+ if (i != 3)
+ continue;
+ break;
+ default:
+ gcc_unreachable ();
+ }
tree t = OMP_CLAUSE_DECL (c);
t = fold_convert (ptr_type_node, t);
gimplify_expr (&t, iseq, NULL, is_gimple_val, fb_rvalue);
@@ -7329,6 +8836,7 @@ lower_depend_clauses (tree *pclauses, gi
}
}
c = build_omp_clause (UNKNOWN_LOCATION, OMP_CLAUSE_DEPEND);
+ OMP_CLAUSE_DEPEND_KIND (c) = OMP_CLAUSE_DEPEND_LAST;
OMP_CLAUSE_DECL (c) = build_fold_addr_expr (array);
OMP_CLAUSE_CHAIN (c) = *pclauses;
*pclauses = c;
@@ -7348,13 +8856,22 @@ lower_omp_taskreg (gimple_stmt_iterator
tree child_fn, t;
gimple *stmt = gsi_stmt (*gsi_p);
gbind *par_bind, *bind, *dep_bind = NULL;
- gimple_seq par_body, olist, ilist, par_olist, par_rlist, par_ilist, new_body;
+ gimple_seq par_body;
location_t loc = gimple_location (stmt);
clauses = gimple_omp_taskreg_clauses (stmt);
- par_bind
- = as_a <gbind *> (gimple_seq_first_stmt (gimple_omp_body (stmt)));
- par_body = gimple_bind_body (par_bind);
+ if (gimple_code (stmt) == GIMPLE_OMP_TASK
+ && gimple_omp_task_taskwait_p (stmt))
+ {
+ par_bind = NULL;
+ par_body = NULL;
+ }
+ else
+ {
+ par_bind
+ = as_a <gbind *> (gimple_seq_first_stmt (gimple_omp_body (stmt)));
+ par_body = gimple_bind_body (par_bind);
+ }
child_fn = ctx->cb.dst_fn;
if (gimple_code (stmt) == GIMPLE_OMP_PARALLEL
&& !gimple_omp_parallel_combined_p (stmt))
@@ -7380,14 +8897,49 @@ lower_omp_taskreg (gimple_stmt_iterator
&dep_ilist, &dep_olist);
}
+ if (gimple_code (stmt) == GIMPLE_OMP_TASK
+ && gimple_omp_task_taskwait_p (stmt))
+ {
+ if (dep_bind)
+ {
+ gsi_replace (gsi_p, dep_bind, true);
+ gimple_bind_add_seq (dep_bind, dep_ilist);
+ gimple_bind_add_stmt (dep_bind, stmt);
+ gimple_bind_add_seq (dep_bind, dep_olist);
+ pop_gimplify_context (dep_bind);
+ }
+ return;
+ }
+
if (ctx->srecord_type)
create_task_copyfn (as_a <gomp_task *> (stmt), ctx);
+ gimple_seq tskred_ilist = NULL;
+ gimple_seq tskred_olist = NULL;
+ if ((is_task_ctx (ctx)
+ && gimple_omp_task_taskloop_p (ctx->stmt)
+ && omp_find_clause (gimple_omp_task_clauses (ctx->stmt),
+ OMP_CLAUSE_REDUCTION))
+ || (is_parallel_ctx (ctx)
+ && omp_find_clause (gimple_omp_parallel_clauses (stmt),
+ OMP_CLAUSE__REDUCTEMP_)))
+ {
+ if (dep_bind == NULL)
+ {
+ push_gimplify_context ();
+ dep_bind = gimple_build_bind (NULL, NULL, make_node (BLOCK));
+ }
+ lower_omp_task_reductions (ctx, is_task_ctx (ctx) ? OMP_TASKLOOP
+ : OMP_PARALLEL,
+ gimple_omp_taskreg_clauses (ctx->stmt),
+ &tskred_ilist, &tskred_olist);
+ }
+
push_gimplify_context ();
- par_olist = NULL;
- par_ilist = NULL;
- par_rlist = NULL;
+ gimple_seq par_olist = NULL;
+ gimple_seq par_ilist = NULL;
+ gimple_seq par_rlist = NULL;
bool phony_construct = gimple_code (stmt) == GIMPLE_OMP_PARALLEL
&& gimple_omp_parallel_grid_phony (as_a <gomp_parallel *> (stmt));
if (phony_construct && ctx->record_type)
@@ -7417,8 +8969,8 @@ lower_omp_taskreg (gimple_stmt_iterator
gimple_omp_taskreg_set_data_arg (stmt, ctx->sender_decl);
}
- olist = NULL;
- ilist = NULL;
+ gimple_seq olist = NULL;
+ gimple_seq ilist = NULL;
lower_send_clauses (clauses, &ilist, &olist, ctx);
lower_send_shared_vars (&ilist, &olist, ctx);
@@ -7433,7 +8985,7 @@ lower_omp_taskreg (gimple_stmt_iterator
/* Once all the expansions are done, sequence all the different
fragments inside gimple_omp_body. */
- new_body = NULL;
+ gimple_seq new_body = NULL;
if (ctx->record_type)
{
@@ -7461,7 +9013,10 @@ lower_omp_taskreg (gimple_stmt_iterator
gimple_omp_set_body (stmt, new_body);
}
- bind = gimple_build_bind (NULL, NULL, gimple_bind_block (par_bind));
+ if (dep_bind && gimple_bind_block (par_bind) == NULL_TREE)
+ bind = gimple_build_bind (NULL, NULL, make_node (BLOCK));
+ else
+ bind = gimple_build_bind (NULL, NULL, gimple_bind_block (par_bind));
gsi_replace (gsi_p, dep_bind ? dep_bind : bind, true);
gimple_bind_add_seq (bind, ilist);
if (!phony_construct)
@@ -7475,7 +9030,9 @@ lower_omp_taskreg (gimple_stmt_iterator
if (dep_bind)
{
gimple_bind_add_seq (dep_bind, dep_ilist);
+ gimple_bind_add_seq (dep_bind, tskred_ilist);
gimple_bind_add_stmt (dep_bind, bind);
+ gimple_bind_add_seq (dep_bind, tskred_olist);
gimple_bind_add_seq (dep_bind, dep_olist);
pop_gimplify_context (dep_bind);
}
@@ -8830,7 +10387,10 @@ lower_omp_1 (gimple_stmt_iterator *gsi_p
case GIMPLE_OMP_TEAMS:
ctx = maybe_lookup_ctx (stmt);
gcc_assert (ctx);
- lower_omp_teams (gsi_p, ctx);
+ if (gimple_omp_teams_host (as_a <gomp_teams *> (stmt)))
+ lower_omp_taskreg (gsi_p, ctx);
+ else
+ lower_omp_teams (gsi_p, ctx);
break;
case GIMPLE_OMP_GRID_BODY:
ctx = maybe_lookup_ctx (stmt);
@@ -282,11 +282,14 @@ unsigned const char omp_clause_num_ops[]
1, /* OMP_CLAUSE_FIRSTPRIVATE */
2, /* OMP_CLAUSE_LASTPRIVATE */
5, /* OMP_CLAUSE_REDUCTION */
+ 5, /* OMP_CLAUSE_TASK_REDUCTION */
+ 5, /* OMP_CLAUSE_IN_REDUCTION */
1, /* OMP_CLAUSE_COPYIN */
1, /* OMP_CLAUSE_COPYPRIVATE */
3, /* OMP_CLAUSE_LINEAR */
2, /* OMP_CLAUSE_ALIGNED */
1, /* OMP_CLAUSE_DEPEND */
+ 1, /* OMP_CLAUSE_NONTEMPORAL */
1, /* OMP_CLAUSE_UNIFORM */
1, /* OMP_CLAUSE_TO_DECLARE */
1, /* OMP_CLAUSE_LINK */
@@ -302,6 +305,7 @@ unsigned const char omp_clause_num_ops[]
0, /* OMP_CLAUSE_AUTO */
0, /* OMP_CLAUSE_SEQ */
1, /* OMP_CLAUSE__LOOPTEMP_ */
+ 1, /* OMP_CLAUSE__REDUCTEMP_ */
1, /* OMP_CLAUSE_IF */
1, /* OMP_CLAUSE_NUM_THREADS */
1, /* OMP_CLAUSE_SCHEDULE */
@@ -355,11 +359,14 @@ const char * const omp_clause_code_name[
"firstprivate",
"lastprivate",
"reduction",
+ "task_reduction",
+ "in_reduction",
"copyin",
"copyprivate",
"linear",
"aligned",
"depend",
+ "nontemporal",
"uniform",
"to",
"link",
@@ -375,6 +382,7 @@ const char * const omp_clause_code_name[
"auto",
"seq",
"_looptemp_",
+ "_reductemp_",
"if",
"num_threads",
"schedule",
@@ -11886,6 +11894,7 @@ walk_tree_1 (tree *tp, walk_tree_fn func
case OMP_CLAUSE_SCHEDULE:
case OMP_CLAUSE_UNIFORM:
case OMP_CLAUSE_DEPEND:
+ case OMP_CLAUSE_NONTEMPORAL:
case OMP_CLAUSE_NUM_TEAMS:
case OMP_CLAUSE_THREAD_LIMIT:
case OMP_CLAUSE_DEVICE:
@@ -11902,6 +11911,7 @@ walk_tree_1 (tree *tp, walk_tree_fn func
case OMP_CLAUSE_USE_DEVICE_PTR:
case OMP_CLAUSE_IS_DEVICE_PTR:
case OMP_CLAUSE__LOOPTEMP_:
+ case OMP_CLAUSE__REDUCTEMP_:
case OMP_CLAUSE__SIMDUID_:
WALK_SUBTREE (OMP_CLAUSE_OPERAND (*tp, 0));
/* FALLTHRU */
@@ -11959,6 +11969,8 @@ walk_tree_1 (tree *tp, walk_tree_fn func
WALK_SUBTREE_TAIL (OMP_CLAUSE_CHAIN (*tp));
case OMP_CLAUSE_REDUCTION:
+ case OMP_CLAUSE_TASK_REDUCTION:
+ case OMP_CLAUSE_IN_REDUCTION:
{
int i;
for (i = 0; i < 5; i++)
@@ -258,6 +258,12 @@ enum omp_clause_code {
reductions. */
OMP_CLAUSE_REDUCTION,
+ /* OpenMP clause: task_reduction (operator:variable_list). */
+ OMP_CLAUSE_TASK_REDUCTION,
+
+ /* OpenMP clause: in_reduction (operator:variable_list). */
+ OMP_CLAUSE_IN_REDUCTION,
+
/* OpenMP clause: copyin (variable_list). */
OMP_CLAUSE_COPYIN,
@@ -273,6 +279,9 @@ enum omp_clause_code {
/* OpenMP clause: depend ({in,out,inout}:variable-list). */
OMP_CLAUSE_DEPEND,
+ /* OpenMP clause: nontemporal (variable-list). */
+ OMP_CLAUSE_NONTEMPORAL,
+
/* OpenMP clause: uniform (argument-list). */
OMP_CLAUSE_UNIFORM,
@@ -331,6 +340,9 @@ enum omp_clause_code {
/* Internal clause: temporary for combined loops expansion. */
OMP_CLAUSE__LOOPTEMP_,
+ /* Internal clause: temporary for task reductions. */
+ OMP_CLAUSE__REDUCTEMP_,
+
/* OpenACC/OpenMP clause: if (scalar-expression). */
OMP_CLAUSE_IF,
@@ -493,6 +505,36 @@ 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)
+};
+
+/* memory-order-clause on OpenMP atomic/flush constructs or
+ argument of atomic_default_mem_order clause. */
+enum omp_memory_order {
+ OMP_MEMORY_ORDER_UNSPECIFIED,
+ OMP_MEMORY_ORDER_RELAXED,
+ OMP_MEMORY_ORDER_ACQUIRE,
+ OMP_MEMORY_ORDER_RELEASE,
+ OMP_MEMORY_ORDER_ACQ_REL,
+ OMP_MEMORY_ORDER_SEQ_CST
+};
+
/* 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. */
@@ -983,6 +1025,9 @@ struct GTY(()) tree_base {
/* Internal function code. */
enum internal_fn ifn;
+ /* OMP_ATOMIC* memory order. */
+ enum omp_memory_order omp_atomic_memory_order;
+
/* The following two fields are used for MEM_REF and TARGET_MEM_REF
expression trees and specify known data non-dependences. For
two memory references in a function they are known to not
@@ -1095,7 +1140,7 @@ struct GTY(()) tree_base {
OMP_CLAUSE_MAP
OMP_CLAUSE_REDUCTION_OMP_ORIG_REF in
- OMP_CLAUSE_REDUCTION
+ OMP_CLAUSE_{,TASK_,IN_}REDUCTION
TRANSACTION_EXPR_RELAXED in
TRANSACTION_EXPR
@@ -1123,9 +1168,6 @@ struct GTY(()) tree_base {
OMP_PARALLEL_COMBINED in
OMP_PARALLEL
- OMP_ATOMIC_SEQ_CST in
- OMP_ATOMIC*
-
OMP_CLAUSE_PRIVATE_OUTER_REF in
OMP_CLAUSE_PRIVATE
@@ -1373,8 +1415,10 @@ enum omp_clause_depend_kind
OMP_CLAUSE_DEPEND_IN,
OMP_CLAUSE_DEPEND_OUT,
OMP_CLAUSE_DEPEND_INOUT,
+ OMP_CLAUSE_DEPEND_MUTEXINOUTSET,
OMP_CLAUSE_DEPEND_SOURCE,
OMP_CLAUSE_DEPEND_SINK,
+ OMP_CLAUSE_DEPEND_DEPOBJ,
OMP_CLAUSE_DEPEND_LAST
};
@@ -1463,6 +1507,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;
@@ -1186,6 +1186,11 @@ DEFTREECODE (OMP_CRITICAL, "omp_critical
Operand 1: OMP_SINGLE_CLAUSES: List of clauses. */
DEFTREECODE (OMP_SINGLE, "omp_single", tcc_statement, 2)
+/* OpenMP - #pragma omp taskgroup
+ Operand 0: OMP_TASKGROUP_BODY: Taskgroup body.
+ Operand 1: OMP_SINGLE_CLAUSES: List of clauses. */
+DEFTREECODE (OMP_TASKGROUP, "omp_taskgroup", tcc_statement, 2)
+
/* OpenMP - #pragma omp section
Operand 0: OMP_SECTION_BODY: Section body. */
DEFTREECODE (OMP_SECTION, "omp_section", tcc_statement, 1)
@@ -1194,10 +1199,6 @@ DEFTREECODE (OMP_SECTION, "omp_section",
Operand 0: OMP_MASTER_BODY: Master section body. */
DEFTREECODE (OMP_MASTER, "omp_master", tcc_statement, 1)
-/* OpenMP - #pragma omp taskgroup
- Operand 0: OMP_TASKGROUP_BODY: Taskgroup body. */
-DEFTREECODE (OMP_TASKGROUP, "omp_taskgroup", tcc_statement, 1)
-
/* OpenACC - #pragma acc cache (variable1 ... variableN)
Operand 0: OACC_CACHE_CLAUSES: List of variables (transformed into
OMP_CLAUSE__CACHE_ clauses). */
@@ -1306,9 +1306,9 @@ extern tree maybe_wrap_with_location (tr
/* Generic accessors for OMP nodes that keep the body as operand 0, and clauses
as operand 1. */
#define OMP_BODY(NODE) \
- TREE_OPERAND (TREE_RANGE_CHECK (NODE, OACC_PARALLEL, OMP_TASKGROUP), 0)
+ TREE_OPERAND (TREE_RANGE_CHECK (NODE, OACC_PARALLEL, OMP_MASTER), 0)
#define OMP_CLAUSES(NODE) \
- TREE_OPERAND (TREE_RANGE_CHECK (NODE, OACC_PARALLEL, OMP_SINGLE), 1)
+ TREE_OPERAND (TREE_RANGE_CHECK (NODE, OACC_PARALLEL, OMP_TASKGROUP), 1)
/* Generic accessors for OMP nodes that keep clauses as operand 0. */
#define OMP_STANDALONE_CLAUSES(NODE) \
@@ -1369,6 +1369,8 @@ extern tree maybe_wrap_with_location (tr
#define OMP_MASTER_BODY(NODE) TREE_OPERAND (OMP_MASTER_CHECK (NODE), 0)
#define OMP_TASKGROUP_BODY(NODE) TREE_OPERAND (OMP_TASKGROUP_CHECK (NODE), 0)
+#define OMP_TASKGROUP_CLAUSES(NODE) \
+ TREE_OPERAND (OMP_TASKGROUP_CHECK (NODE), 1)
#define OMP_ORDERED_BODY(NODE) TREE_OPERAND (OMP_ORDERED_CHECK (NODE), 0)
#define OMP_ORDERED_CLAUSES(NODE) TREE_OPERAND (OMP_ORDERED_CHECK (NODE), 1)
@@ -1406,7 +1408,7 @@ extern tree maybe_wrap_with_location (tr
#define OMP_CLAUSE_DECL(NODE) \
OMP_CLAUSE_OPERAND (OMP_CLAUSE_RANGE_CHECK (OMP_CLAUSE_CHECK (NODE), \
OMP_CLAUSE_PRIVATE, \
- OMP_CLAUSE__LOOPTEMP_), 0)
+ OMP_CLAUSE__REDUCTEMP_), 0)
#define OMP_CLAUSE_HAS_LOCATION(NODE) \
(LOCATION_LOCUS ((OMP_CLAUSE_CHECK (NODE))->omp_clause.locus) \
!= UNKNOWN_LOCATION)
@@ -1432,11 +1434,10 @@ extern tree maybe_wrap_with_location (tr
#define OMP_TARGET_COMBINED(NODE) \
(OMP_TARGET_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) \
+/* Memory order for OMP_ATOMIC*. */
+#define OMP_ATOMIC_MEMORY_ORDER(NODE) \
(TREE_RANGE_CHECK (NODE, OMP_ATOMIC, \
- OMP_ATOMIC_CAPTURE_NEW)->base.private_flag)
+ OMP_ATOMIC_CAPTURE_NEW)->base.u.omp_atomic_memory_order)
/* True on a PRIVATE clause if its decl is kept around for debugging
information only and its DECL_VALUE_EXPR is supposed to point
@@ -1459,6 +1460,11 @@ extern tree maybe_wrap_with_location (tr
#define OMP_CLAUSE_FIRSTPRIVATE_IMPLICIT(NODE) \
(OMP_CLAUSE_SUBCODE_CHECK (NODE, OMP_CLAUSE_FIRSTPRIVATE)->base.public_flag)
+/* True on a FIRSTPRIVATE clause if only the reference and not what it refers
+ to should be firstprivatized. */
+#define OMP_CLAUSE_FIRSTPRIVATE_NO_REFERENCE(NODE) \
+ TREE_PRIVATE (OMP_CLAUSE_SUBCODE_CHECK (NODE, OMP_CLAUSE_FIRSTPRIVATE))
+
/* True on a LASTPRIVATE clause if a FIRSTPRIVATE clause for the same
decl is present in the chain. */
#define OMP_CLAUSE_LASTPRIVATE_FIRSTPRIVATE(NODE) \
@@ -1476,6 +1482,10 @@ extern tree maybe_wrap_with_location (tr
#define OMP_CLAUSE_LASTPRIVATE_TASKLOOP_IV(NODE) \
TREE_PROTECTED (OMP_CLAUSE_SUBCODE_CHECK (NODE, OMP_CLAUSE_LASTPRIVATE))
+/* True if a LASTPRIVATE clause has CONDITIONAL: modifier. */
+#define OMP_CLAUSE_LASTPRIVATE_CONDITIONAL(NODE) \
+ TREE_PRIVATE (OMP_CLAUSE_SUBCODE_CHECK (NODE, OMP_CLAUSE_LASTPRIVATE))
+
/* True on a SHARED clause if a FIRSTPRIVATE clause for the same
decl is present in the chain (this can happen only for taskloop
with FIRSTPRIVATE/LASTPRIVATE on it originally. */
@@ -1579,24 +1589,38 @@ extern tree maybe_wrap_with_location (tr
OMP_CLAUSE_OPERAND (OMP_CLAUSE_SUBCODE_CHECK (NODE, OMP_CLAUSE_ORDERED), 0)
#define OMP_CLAUSE_REDUCTION_CODE(NODE) \
- (OMP_CLAUSE_SUBCODE_CHECK (NODE, OMP_CLAUSE_REDUCTION)->omp_clause.subcode.reduction_code)
+ (OMP_CLAUSE_RANGE_CHECK (NODE, OMP_CLAUSE_REDUCTION, \
+ OMP_CLAUSE_IN_REDUCTION)->omp_clause.subcode.reduction_code)
#define OMP_CLAUSE_REDUCTION_INIT(NODE) \
- OMP_CLAUSE_OPERAND (OMP_CLAUSE_SUBCODE_CHECK (NODE, OMP_CLAUSE_REDUCTION), 1)
+ OMP_CLAUSE_OPERAND (OMP_CLAUSE_RANGE_CHECK (NODE, OMP_CLAUSE_REDUCTION, \
+ OMP_CLAUSE_IN_REDUCTION), 1)
#define OMP_CLAUSE_REDUCTION_MERGE(NODE) \
- OMP_CLAUSE_OPERAND (OMP_CLAUSE_SUBCODE_CHECK (NODE, OMP_CLAUSE_REDUCTION), 2)
+ OMP_CLAUSE_OPERAND (OMP_CLAUSE_RANGE_CHECK (NODE, OMP_CLAUSE_REDUCTION, \
+ OMP_CLAUSE_IN_REDUCTION), 2)
#define OMP_CLAUSE_REDUCTION_GIMPLE_INIT(NODE) \
(OMP_CLAUSE_CHECK (NODE))->omp_clause.gimple_reduction_init
#define OMP_CLAUSE_REDUCTION_GIMPLE_MERGE(NODE) \
(OMP_CLAUSE_CHECK (NODE))->omp_clause.gimple_reduction_merge
#define OMP_CLAUSE_REDUCTION_PLACEHOLDER(NODE) \
- OMP_CLAUSE_OPERAND (OMP_CLAUSE_SUBCODE_CHECK (NODE, OMP_CLAUSE_REDUCTION), 3)
+ OMP_CLAUSE_OPERAND (OMP_CLAUSE_RANGE_CHECK (NODE, OMP_CLAUSE_REDUCTION, \
+ OMP_CLAUSE_IN_REDUCTION), 3)
#define OMP_CLAUSE_REDUCTION_DECL_PLACEHOLDER(NODE) \
- OMP_CLAUSE_OPERAND (OMP_CLAUSE_SUBCODE_CHECK (NODE, OMP_CLAUSE_REDUCTION), 4)
+ OMP_CLAUSE_OPERAND (OMP_CLAUSE_RANGE_CHECK (NODE, OMP_CLAUSE_REDUCTION, \
+ OMP_CLAUSE_IN_REDUCTION), 4)
/* True if a REDUCTION clause may reference the original list item (omp_orig)
in its OMP_CLAUSE_REDUCTION_{,GIMPLE_}INIT. */
#define OMP_CLAUSE_REDUCTION_OMP_ORIG_REF(NODE) \
- (OMP_CLAUSE_SUBCODE_CHECK (NODE, OMP_CLAUSE_REDUCTION)->base.public_flag)
+ (OMP_CLAUSE_RANGE_CHECK (NODE, OMP_CLAUSE_REDUCTION, \
+ OMP_CLAUSE_IN_REDUCTION)->base.public_flag)
+
+/* True if a REDUCTION clause has task reduction-modifier. */
+#define OMP_CLAUSE_REDUCTION_TASK(NODE) \
+ TREE_PROTECTED (OMP_CLAUSE_SUBCODE_CHECK (NODE, OMP_CLAUSE_REDUCTION))
+
+/* True if a REDUCTION clause has inscan reduction-modifier. */
+#define OMP_CLAUSE_REDUCTION_INSCAN(NODE) \
+ TREE_PRIVATE (OMP_CLAUSE_SUBCODE_CHECK (NODE, OMP_CLAUSE_REDUCTION))
/* True if a LINEAR clause doesn't need copy in. True for iterator vars which
are always initialized inside of the loop construct, false otherwise. */
@@ -1665,6 +1689,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) \
@@ -1515,7 +1515,8 @@ remap_gimple_stmt (gimple *stmt, copy_bo
case GIMPLE_OMP_TASKGROUP:
s1 = remap_gimple_seq (gimple_omp_body (stmt), id);
- copy = gimple_build_omp_taskgroup (s1);
+ copy = gimple_build_omp_taskgroup
+ (s1, gimple_omp_taskgroup_clauses (stmt));
break;
case GIMPLE_OMP_ORDERED:
@@ -1314,6 +1314,8 @@ convert_nonlocal_omp_clauses (tree *pcla
convert_nonlocal_reference_op
(&OMP_CLAUSE_ALIGNED_ALIGNMENT (clause), &dummy, wi);
}
+ /* FALLTHRU */
+ case OMP_CLAUSE_NONTEMPORAL:
/* Like do_decl_clause, but don't add any suppression. */
decl = OMP_CLAUSE_DECL (clause);
if (VAR_P (decl)
@@ -1322,8 +1324,7 @@ convert_nonlocal_omp_clauses (tree *pcla
if (decl_function_context (decl) != info->context)
{
OMP_CLAUSE_DECL (clause) = get_nonlocal_debug_decl (info, decl);
- if (OMP_CLAUSE_CODE (clause) != OMP_CLAUSE_PRIVATE)
- need_chain = true;
+ need_chain = true;
}
break;
@@ -1365,6 +1366,7 @@ convert_nonlocal_omp_clauses (tree *pcla
/* The following clauses are only added during OMP lowering; nested
function decomposition happens before that. */
case OMP_CLAUSE__LOOPTEMP_:
+ case OMP_CLAUSE__REDUCTEMP_:
case OMP_CLAUSE__SIMDUID_:
case OMP_CLAUSE__GRIDDIM_:
/* Anything else. */
@@ -2000,6 +2002,8 @@ convert_local_omp_clauses (tree *pclause
convert_local_reference_op
(&OMP_CLAUSE_ALIGNED_ALIGNMENT (clause), &dummy, wi);
}
+ /* FALLTHRU */
+ case OMP_CLAUSE_NONTEMPORAL:
/* Like do_decl_clause, but don't add any suppression. */
decl = OMP_CLAUSE_DECL (clause);
if (VAR_P (decl)
@@ -2056,6 +2060,7 @@ convert_local_omp_clauses (tree *pclause
/* The following clauses are only added during OMP lowering; nested
function decomposition happens before that. */
case OMP_CLAUSE__LOOPTEMP_:
+ case OMP_CLAUSE__REDUCTEMP_:
case OMP_CLAUSE__SIMDUID_:
case OMP_CLAUSE__GRIDDIM_:
/* Anything else. */
@@ -1130,7 +1130,8 @@ create_call_for_reduction_1 (reduction_i
tmp_load = create_tmp_var (TREE_TYPE (TREE_TYPE (addr)));
tmp_load = make_ssa_name (tmp_load);
- load = gimple_build_omp_atomic_load (tmp_load, addr);
+ load = gimple_build_omp_atomic_load (tmp_load, addr,
+ OMP_MEMORY_ORDER_RELAXED);
SSA_NAME_DEF_STMT (tmp_load) = load;
gsi = gsi_start_bb (new_bb);
gsi_insert_after (&gsi, load, GSI_NEW_STMT);
@@ -1146,7 +1147,9 @@ create_call_for_reduction_1 (reduction_i
name = force_gimple_operand_gsi (&gsi, x, true, NULL_TREE, true,
GSI_CONTINUE_LINKING);
- gsi_insert_after (&gsi, gimple_build_omp_atomic_store (name), GSI_NEW_STMT);
+ gimple *store = gimple_build_omp_atomic_store (name,
+ OMP_MEMORY_ORDER_RELAXED);
+ gsi_insert_after (&gsi, store, GSI_NEW_STMT);
return 1;
}
@@ -392,6 +392,31 @@ dump_array_domain (pretty_printer *pp, t
}
+/* Dump OpenMP iterators ITER. */
+
+static void
+dump_omp_iterators (pretty_printer *pp, tree iter, int spc, dump_flags_t flags)
+{
+ pp_string (pp, "iterator(");
+ for (tree it = iter; it; it = TREE_CHAIN (it))
+ {
+ if (it != iter)
+ pp_string (pp, ", ");
+ dump_generic_node (pp, TREE_TYPE (TREE_VEC_ELT (it, 0)), spc, flags,
+ false);
+ pp_space (pp);
+ dump_generic_node (pp, TREE_VEC_ELT (it, 0), spc, flags, false);
+ pp_equal (pp);
+ dump_generic_node (pp, TREE_VEC_ELT (it, 1), spc, flags, false);
+ pp_colon (pp);
+ dump_generic_node (pp, TREE_VEC_ELT (it, 2), spc, flags, false);
+ pp_colon (pp);
+ dump_generic_node (pp, TREE_VEC_ELT (it, 3), spc, flags, false);
+ }
+ pp_right_paren (pp);
+}
+
+
/* Dump OpenMP clause CLAUSE. PP, CLAUSE, SPC and FLAGS are as in
dump_generic_node. */
@@ -413,7 +438,13 @@ dump_omp_clause (pretty_printer *pp, tre
goto print_remap;
case OMP_CLAUSE_LASTPRIVATE:
name = "lastprivate";
- goto print_remap;
+ if (!OMP_CLAUSE_LASTPRIVATE_CONDITIONAL (clause))
+ goto print_remap;
+ pp_string (pp, "lastprivate(conditional:");
+ dump_generic_node (pp, OMP_CLAUSE_DECL (clause),
+ spc, flags, false);
+ pp_right_paren (pp);
+ break;
case OMP_CLAUSE_COPYIN:
name = "copyin";
goto print_remap;
@@ -432,12 +463,18 @@ dump_omp_clause (pretty_printer *pp, tre
case OMP_CLAUSE__LOOPTEMP_:
name = "_looptemp_";
goto print_remap;
+ case OMP_CLAUSE__REDUCTEMP_:
+ name = "_reductemp_";
+ goto print_remap;
case OMP_CLAUSE_TO_DECLARE:
name = "to";
goto print_remap;
case OMP_CLAUSE_LINK:
name = "link";
goto print_remap;
+ case OMP_CLAUSE_NONTEMPORAL:
+ name = "nontemporal";
+ goto print_remap;
print_remap:
pp_string (pp, name);
pp_left_paren (pp);
@@ -446,8 +483,20 @@ dump_omp_clause (pretty_printer *pp, tre
pp_right_paren (pp);
break;
+ case OMP_CLAUSE_TASK_REDUCTION:
+ case OMP_CLAUSE_IN_REDUCTION:
+ pp_string (pp, OMP_CLAUSE_CODE (clause) == OMP_CLAUSE_IN_REDUCTION
+ ? "in_" : "task_");
+ /* FALLTHRU */
case OMP_CLAUSE_REDUCTION:
pp_string (pp, "reduction(");
+ if (OMP_CLAUSE_CODE (clause) == OMP_CLAUSE_REDUCTION)
+ {
+ if (OMP_CLAUSE_REDUCTION_TASK (clause))
+ pp_string (pp, "task,");
+ else if (OMP_CLAUSE_REDUCTION_INSCAN (clause))
+ pp_string (pp, "inscan,");
+ }
if (OMP_CLAUSE_REDUCTION_CODE (clause) != ERROR_MARK)
{
pp_string (pp,
@@ -464,7 +513,9 @@ dump_omp_clause (pretty_printer *pp, tre
switch (OMP_CLAUSE_IF_MODIFIER (clause))
{
case ERROR_MARK: break;
+ case VOID_CST: pp_string (pp, "cancel:"); break;
case OMP_PARALLEL: pp_string (pp, "parallel:"); break;
+ case OMP_SIMD: pp_string (pp, "simd:"); break;
case OMP_TASK: pp_string (pp, "task:"); break;
case OMP_TASKLOOP: pp_string (pp, "taskloop:"); break;
case OMP_TARGET_DATA: pp_string (pp, "target data:"); break;
@@ -643,18 +694,27 @@ dump_omp_clause (pretty_printer *pp, tre
pp_string (pp, "depend(");
switch (OMP_CLAUSE_DEPEND_KIND (clause))
{
+ case OMP_CLAUSE_DEPEND_DEPOBJ:
+ name = "depobj";
+ break;
case OMP_CLAUSE_DEPEND_IN:
- pp_string (pp, "in");
+ name = "in";
break;
case OMP_CLAUSE_DEPEND_OUT:
- pp_string (pp, "out");
+ name = "out";
break;
case OMP_CLAUSE_DEPEND_INOUT:
- pp_string (pp, "inout");
+ name = "inout";
+ break;
+ case OMP_CLAUSE_DEPEND_MUTEXINOUTSET:
+ name = "mutexinoutset";
break;
case OMP_CLAUSE_DEPEND_SOURCE:
pp_string (pp, "source)");
return;
+ case OMP_CLAUSE_DEPEND_LAST:
+ name = "__internal__";
+ break;
case OMP_CLAUSE_DEPEND_SINK:
pp_string (pp, "sink:");
for (tree t = OMP_CLAUSE_DECL (clause); t; t = TREE_CHAIN (t))
@@ -680,10 +740,21 @@ dump_omp_clause (pretty_printer *pp, tre
default:
gcc_unreachable ();
}
- pp_colon (pp);
- dump_generic_node (pp, OMP_CLAUSE_DECL (clause),
- spc, flags, false);
- pp_right_paren (pp);
+ {
+ tree t = OMP_CLAUSE_DECL (clause);
+ if (TREE_CODE (t) == TREE_LIST
+ && TREE_PURPOSE (t)
+ && TREE_CODE (TREE_PURPOSE (t)) == TREE_VEC)
+ {
+ dump_omp_iterators (pp, TREE_PURPOSE (t), spc, flags);
+ pp_colon (pp);
+ t = TREE_VALUE (t);
+ }
+ pp_string (pp, name);
+ pp_colon (pp);
+ dump_generic_node (pp, t, spc, flags, false);
+ pp_right_paren (pp);
+ }
break;
case OMP_CLAUSE_MAP:
@@ -900,7 +971,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_:
@@ -1218,6 +1335,34 @@ dump_block_node (pretty_printer *pp, tre
}
}
+/* Dump #pragma omp atomic memory order clause. */
+
+void
+dump_omp_atomic_memory_order (pretty_printer *pp, enum omp_memory_order mo)
+{
+ switch (mo)
+ {
+ case OMP_MEMORY_ORDER_RELAXED:
+ pp_string (pp, " relaxed");
+ break;
+ case OMP_MEMORY_ORDER_SEQ_CST:
+ pp_string (pp, " seq_cst");
+ break;
+ case OMP_MEMORY_ORDER_ACQ_REL:
+ pp_string (pp, " acq_rel");
+ break;
+ case OMP_MEMORY_ORDER_ACQUIRE:
+ pp_string (pp, " acquire");
+ break;
+ case OMP_MEMORY_ORDER_RELEASE:
+ pp_string (pp, " release");
+ break;
+ case OMP_MEMORY_ORDER_UNSPECIFIED:
+ break;
+ default:
+ gcc_unreachable ();
+ }
+}
/* Dump the node NODE on the pretty_printer PP, SPC spaces of
indent. FLAGS specifies details to show in the dump (see TDF_* in
@@ -2991,7 +3136,8 @@ dump_generic_node (pretty_printer *pp, t
break;
case OMP_TASK:
- pp_string (pp, "#pragma omp task");
+ pp_string (pp, OMP_TASK_BODY (node) ? "#pragma omp task"
+ : "#pragma omp taskwait");
dump_omp_clauses (pp, OMP_TASK_CLAUSES (node), spc, flags);
goto dump_omp_body;
@@ -3122,6 +3268,7 @@ dump_generic_node (pretty_printer *pp, t
case OMP_TASKGROUP:
pp_string (pp, "#pragma omp taskgroup");
+ dump_omp_clauses (pp, OMP_TASKGROUP_CLAUSES (node), spc, flags);
goto dump_omp_body;
case OMP_ORDERED:
@@ -3144,8 +3291,7 @@ dump_generic_node (pretty_printer *pp, t
case OMP_ATOMIC:
pp_string (pp, "#pragma omp atomic");
- if (OMP_ATOMIC_SEQ_CST (node))
- pp_string (pp, " seq_cst");
+ dump_omp_atomic_memory_order (pp, OMP_ATOMIC_MEMORY_ORDER (node));
newline_and_indent (pp, spc + 2);
dump_generic_node (pp, TREE_OPERAND (node, 0), spc, flags, false);
pp_space (pp);
@@ -3156,8 +3302,7 @@ dump_generic_node (pretty_printer *pp, t
case OMP_ATOMIC_READ:
pp_string (pp, "#pragma omp atomic read");
- if (OMP_ATOMIC_SEQ_CST (node))
- pp_string (pp, " seq_cst");
+ dump_omp_atomic_memory_order (pp, OMP_ATOMIC_MEMORY_ORDER (node));
newline_and_indent (pp, spc + 2);
dump_generic_node (pp, TREE_OPERAND (node, 0), spc, flags, false);
pp_space (pp);
@@ -3166,8 +3311,7 @@ dump_generic_node (pretty_printer *pp, t
case OMP_ATOMIC_CAPTURE_OLD:
case OMP_ATOMIC_CAPTURE_NEW:
pp_string (pp, "#pragma omp atomic capture");
- if (OMP_ATOMIC_SEQ_CST (node))
- pp_string (pp, " seq_cst");
+ dump_omp_atomic_memory_order (pp, OMP_ATOMIC_MEMORY_ORDER (node));
newline_and_indent (pp, spc + 2);
dump_generic_node (pp, TREE_OPERAND (node, 0), spc, flags, false);
pp_space (pp);
@@ -40,6 +40,8 @@ extern void print_generic_stmt_indented
extern void print_generic_expr (FILE *, tree, dump_flags_t = TDF_NONE);
extern char *print_generic_expr_to_str (tree);
extern void dump_omp_clauses (pretty_printer *, tree, int, dump_flags_t);
+extern void dump_omp_atomic_memory_order (pretty_printer *,
+ enum omp_memory_order);
extern int dump_generic_node (pretty_printer *, tree, int, dump_flags_t, bool);
extern void print_declaration (pretty_printer *, tree, int, dump_flags_t);
extern int op_code_prio (enum tree_code);
@@ -447,6 +447,8 @@ unpack_ts_omp_clause_value_fields (struc
OMP_CLAUSE_PROC_BIND_LAST);
break;
case OMP_CLAUSE_REDUCTION:
+ case OMP_CLAUSE_TASK_REDUCTION:
+ case OMP_CLAUSE_IN_REDUCTION:
OMP_CLAUSE_REDUCTION_CODE (expr)
= bp_unpack_enum (bp, tree_code, MAX_TREE_CODES);
break;
@@ -395,6 +395,8 @@ pack_ts_omp_clause_value_fields (struct
OMP_CLAUSE_PROC_BIND_KIND (expr));
break;
case OMP_CLAUSE_REDUCTION:
+ case OMP_CLAUSE_TASK_REDUCTION:
+ case OMP_CLAUSE_IN_REDUCTION:
bp_pack_enum (bp, tree_code, MAX_TREE_CODES,
OMP_CLAUSE_REDUCTION_CODE (expr));
break;
@@ -836,12 +838,18 @@ write_ts_omp_clause_tree_pointers (struc
int i;
for (i = 0; i < omp_clause_num_ops[OMP_CLAUSE_CODE (expr)]; i++)
stream_write_tree (ob, OMP_CLAUSE_OPERAND (expr, i), ref_p);
- if (OMP_CLAUSE_CODE (expr) == OMP_CLAUSE_REDUCTION)
+ switch (OMP_CLAUSE_CODE (expr))
{
+ case OMP_CLAUSE_REDUCTION:
+ case OMP_CLAUSE_TASK_REDUCTION:
+ case OMP_CLAUSE_IN_REDUCTION:
/* We don't stream these right now, handle it if streaming
of them is needed. */
gcc_assert (OMP_CLAUSE_REDUCTION_GIMPLE_INIT (expr) == NULL);
gcc_assert (OMP_CLAUSE_REDUCTION_GIMPLE_MERGE (expr) == NULL);
+ break;
+ default:
+ break;
}
stream_write_tree (ob, OMP_CLAUSE_CHAIN (expr), ref_p);
}
@@ -189,6 +189,7 @@ enum gomp_map_kind
#define GOMP_TASK_FLAG_GRAINSIZE (1 << 9)
#define GOMP_TASK_FLAG_IF (1 << 10)
#define GOMP_TASK_FLAG_NOGROUP (1 << 11)
+#define GOMP_TASK_FLAG_REDUCTION (1 << 12)
/* GOMP_target{_ext,update_ext,enter_exit_data} flags argument. */
#define GOMP_TARGET_FLAG_NOWAIT (1 << 0)
@@ -251,6 +252,12 @@ enum gomp_map_kind
at most and shifted by this many bits. */
#define GOMP_TARGET_ARG_VALUE_SHIFT 16
+/* Dependence types in omp_depend_t objects. */
+#define GOMP_DEPEND_IN 1
+#define GOMP_DEPEND_OUT 2
+#define GOMP_DEPEND_INOUT 3
+#define GOMP_DEPEND_MUTEXINOUTSET 4
+
/* HSA specific data structures. */
/* Identifiers of device-specific target arguments. */