diff --git a/gcc/Makefile.in b/gcc/Makefile.in index 1164554e6d6..28e29fab93d 100644 --- a/gcc/Makefile.in +++ b/gcc/Makefile.in @@ -1505,6 +1505,7 @@ OBJS = \ omp-general.o \ omp-low.o \ omp-oacc-kernels-decompose.o \ + omp-expand-metadirective.o \ omp-simd-clone.o \ opt-problem.o \ optabs.o \ diff --git a/gcc/c-family/c-pragma.c b/gcc/c-family/c-pragma.c index 4f8e8e0128c..01dc1e6d9c0 100644 --- a/gcc/c-family/c-pragma.c +++ b/gcc/c-family/c-pragma.c @@ -1312,12 +1312,14 @@ static const struct omp_pragma_def omp_pragmas[] = { { "allocate", PRAGMA_OMP_ALLOCATE }, { "atomic", PRAGMA_OMP_ATOMIC }, { "barrier", PRAGMA_OMP_BARRIER }, + { "begin", PRAGMA_OMP_BEGIN }, { "cancel", PRAGMA_OMP_CANCEL }, { "cancellation", PRAGMA_OMP_CANCELLATION_POINT }, { "critical", PRAGMA_OMP_CRITICAL }, { "depobj", PRAGMA_OMP_DEPOBJ }, - { "end", PRAGMA_OMP_END_DECLARE_TARGET }, + { "end", PRAGMA_OMP_END }, { "flush", PRAGMA_OMP_FLUSH }, + { "metadirective", PRAGMA_OMP_METADIRECTIVE }, { "requires", PRAGMA_OMP_REQUIRES }, { "section", PRAGMA_OMP_SECTION }, { "sections", PRAGMA_OMP_SECTIONS }, @@ -1387,6 +1389,41 @@ c_pp_lookup_pragma (unsigned int id, const char **space, const char **name) gcc_unreachable (); } +static int +c_pp_lookup_pragma_by_name_1 (const void *name, const void *elem) +{ + const struct omp_pragma_def *pragma_def + = (const struct omp_pragma_def *) elem; + + return strcmp ((const char *) name, pragma_def->name); +} + +enum pragma_kind +c_pp_lookup_pragma_by_name (const char *name) +{ + const int n_omp_pragmas = sizeof (omp_pragmas) / sizeof (*omp_pragmas); + const int n_omp_pragmas_simd = sizeof (omp_pragmas_simd) + / sizeof (*omp_pragmas_simd); + + void *result = bsearch (name, omp_pragmas, n_omp_pragmas, + sizeof (*omp_pragmas), + c_pp_lookup_pragma_by_name_1); + if (!result) + result = bsearch (name, omp_pragmas_simd, n_omp_pragmas_simd, + sizeof (*omp_pragmas_simd), + c_pp_lookup_pragma_by_name_1); + + if (result) + { + const struct omp_pragma_def *def + = (const struct omp_pragma_def *) result; + + return (enum pragma_kind) def->id; + } + + return PRAGMA_NONE; +} + /* Front-end wrappers for pragma registration to avoid dragging cpplib.h in almost everywhere. */ diff --git a/gcc/c-family/c-pragma.h b/gcc/c-family/c-pragma.h index 6c34ffa5be4..6d4698d41ba 100644 --- a/gcc/c-family/c-pragma.h +++ b/gcc/c-family/c-pragma.h @@ -45,17 +45,19 @@ enum pragma_kind { PRAGMA_OMP_ALLOCATE, PRAGMA_OMP_ATOMIC, PRAGMA_OMP_BARRIER, + PRAGMA_OMP_BEGIN, PRAGMA_OMP_CANCEL, PRAGMA_OMP_CANCELLATION_POINT, PRAGMA_OMP_CRITICAL, PRAGMA_OMP_DECLARE, PRAGMA_OMP_DEPOBJ, PRAGMA_OMP_DISTRIBUTE, - PRAGMA_OMP_END_DECLARE_TARGET, + PRAGMA_OMP_END, PRAGMA_OMP_FLUSH, PRAGMA_OMP_FOR, PRAGMA_OMP_LOOP, PRAGMA_OMP_MASTER, + PRAGMA_OMP_METADIRECTIVE, PRAGMA_OMP_ORDERED, PRAGMA_OMP_PARALLEL, PRAGMA_OMP_REQUIRES, @@ -252,6 +254,7 @@ extern enum cpp_ttype c_lex_with_flags (tree *, location_t *, unsigned char *, int); extern void c_pp_lookup_pragma (unsigned int, const char **, const char **); +extern enum pragma_kind c_pp_lookup_pragma_by_name (const char *); extern GTY(()) tree pragma_extern_prefix; diff --git a/gcc/c/c-parser.c b/gcc/c/c-parser.c index b9930d487fd..fa807530ca7 100644 --- a/gcc/c/c-parser.c +++ b/gcc/c/c-parser.c @@ -1583,8 +1583,12 @@ enum pragma_context { pragma_external, pragma_struct, pragma_param, static bool c_parser_pragma (c_parser *, enum pragma_context, bool *); static void c_parser_omp_cancellation_point (c_parser *, enum pragma_context); static bool c_parser_omp_target (c_parser *, enum pragma_context, bool *); +static void c_parser_omp_begin (c_parser *, bool *); +static void c_parser_omp_end (c_parser *); static void c_parser_omp_end_declare_target (c_parser *); static void c_parser_omp_declare (c_parser *, enum pragma_context); +static tree c_parser_omp_metadirective (location_t, c_parser *, char *, + omp_clause_mask, tree *, bool *, bool); static void c_parser_omp_requires (c_parser *); static bool c_parser_omp_ordered (c_parser *, enum pragma_context, bool *); static void c_parser_oacc_routine (c_parser *, enum pragma_context); @@ -12402,8 +12406,12 @@ c_parser_pragma (c_parser *parser, enum pragma_context context, bool *if_p) case PRAGMA_OMP_TARGET: return c_parser_omp_target (parser, context, if_p); - case PRAGMA_OMP_END_DECLARE_TARGET: - c_parser_omp_end_declare_target (parser); + case PRAGMA_OMP_BEGIN: + c_parser_omp_begin (parser, if_p); + return false; + + case PRAGMA_OMP_END: + c_parser_omp_end (parser); return false; case PRAGMA_OMP_SCAN: @@ -18195,6 +18203,7 @@ c_parser_omp_for_loop (location_t loc, c_parser *parser, enum tree_code code, location_t for_loc; bool tiling = false; bool inscan = false; + vec *for_block = make_tree_vector (); for (cl = clauses; cl; cl = OMP_CLAUSE_CHAIN (cl)) @@ -20934,6 +20943,60 @@ c_parser_omp_end_declare_target (c_parser *parser) current_omp_declare_target_attribute--; } +static void +c_parser_omp_begin (c_parser *parser, bool *if_p) +{ + location_t loc = c_parser_peek_token (parser)->location; + c_parser_consume_pragma(parser); + if (c_parser_peek_token (parser)->type == CPP_NAME) + { + const char *p = IDENTIFIER_POINTER (c_parser_peek_token (parser)->value); + + if (strcmp (p, "metadirective") == 0) + { + char p_name[sizeof "#pragma omp teams distribute parallel for simd"]; + omp_clause_mask mask (0); + + c_parser_consume_token (parser); + c_parser_omp_metadirective (loc, parser, p_name, mask, NULL, if_p, + true); + return; + } + } + + error_at (loc, "expected %"); + c_parser_skip_to_pragma_eol (parser); +} + +static void +c_parser_omp_end (c_parser *parser) +{ + location_t loc = c_parser_peek_token (parser)->location; + + if (c_parser_peek_2nd_token (parser)->type == CPP_NAME) + { + const char *p + = IDENTIFIER_POINTER (c_parser_peek_2nd_token (parser)->value); + + if (strcmp (p, "declare") == 0) + { + c_parser_omp_end_declare_target (parser); + return; + } + else if (strcmp (p, "metadirective") == 0) + { + /* The pragma 'omp end metadirective' should have been consumed + when processing the metadirective. */ + error_at (loc, "%<#pragma omp end metadirective%> without " + "corresponding %<#pragma omp begin metadirective%>"); + } + } + else + error_at (loc, "expected % or %"); + + c_parser_skip_to_pragma_eol (parser); +} + /* OpenMP 4.0 #pragma omp declare reduction (reduction-id : typename-list : expression) \ @@ -21607,6 +21670,295 @@ c_parser_omp_taskloop (location_t loc, c_parser *parser, return ret; } +/* OpenMP 5.0: + + # pragma omp metadirective [clause[, clause]] + + # pragma omp begin metadirective [clause[, clause]] + # pragma omp end metadirective +*/ + +static tree +c_parser_omp_metadirective (location_t loc, c_parser *parser, + char *p_name, omp_clause_mask mask, tree *cclauses, + bool *if_p, + bool begin_end_p) +{ + tree ret; + bool all_selectors_resolveable = true; + auto_vec > directive_tokens; + auto_vec ctxs; + bool default_seen = false; + + ret = make_node (OMP_METADIRECTIVE); + SET_EXPR_LOCATION (ret, loc); + TREE_TYPE (ret) = void_type_node; + OMP_METADIRECTIVE_CLAUSES (ret) = NULL_TREE; + strcat (p_name, " metadirective"); + + while (c_parser_next_token_is_not (parser, CPP_PRAGMA_EOL)) + { + if (c_parser_next_token_is_not (parser, CPP_NAME) + && c_parser_next_token_is_not (parser, CPP_KEYWORD)) + { + c_parser_error (parser, "expected % or %"); + return NULL_TREE; + } + + location_t match_loc = c_parser_peek_token (parser)->location; + const char *p = IDENTIFIER_POINTER (c_parser_peek_token (parser)->value); + c_parser_consume_token (parser); + bool default_p = strcmp (p, "default") == 0; + if (default_p) + { + if (default_seen) + { + c_parser_error (parser, "there can only be one default clause"); + return NULL_TREE; + } + else + default_seen = true; + } + if (strcmp (p, "when") == 0 || default_p) + { + matching_parens parens; + tree ctx = NULL_TREE; + bool skip = false; + if (!parens.require_open (parser)) + return error_mark_node; + + if (!default_p) + { + ctx = c_parser_omp_context_selector_specification (parser, + NULL_TREE); + if (ctx == error_mark_node) + return NULL_TREE; + ctx = c_omp_check_context_selector (match_loc, ctx); + if (ctx == error_mark_node) + return NULL_TREE; + + switch (omp_context_selector_matches (ctx, true)) + { + case -1: + all_selectors_resolveable = false; + break; + case 1: + break; + case 0: + /* Remove the selector from further consideration. */ + skip = true; + break; + } + + if (c_parser_next_token_is_not (parser, CPP_COLON)) + { + c_parser_error (parser, "expected colon"); + return NULL_TREE; + } + c_parser_consume_token (parser); + } + + /* Read in the directive type and create a dummy pragma token for + it. */ + c_token *token = c_parser_peek_token (parser); + if (token->type != CPP_NAME) + { + c_parser_error (parser, "expected directive name"); + return NULL_TREE; + } + + location_t loc = c_parser_peek_token (parser)->location; + const char *p + = IDENTIFIER_POINTER (c_parser_peek_token (parser)->value); + enum pragma_kind p_kind = c_pp_lookup_pragma_by_name (p); + + c_parser_consume_token (parser); + if (p_kind == PRAGMA_NONE) + { + c_parser_error (parser, "unknown directive name"); + return NULL_TREE; + } + + if (!skip) + { + c_token pragma_token; + pragma_token.type = CPP_PRAGMA; + pragma_token.location = loc; + pragma_token.pragma_kind = p_kind; + + directive_tokens.safe_push (auto_vec ()); + directive_tokens.last ().safe_push (pragma_token); + + ctxs.safe_push (ctx); + } + + /* Read in tokens for the directive clauses. */ + auto_vec *tokens = skip ? NULL : &directive_tokens.last (); + int nesting_depth = 0; + while (1) + { + c_token *token = c_parser_peek_token (parser); + switch (token->type) + { + case CPP_EOF: + case CPP_PRAGMA_EOL: + break; + case CPP_OPEN_PAREN: + ++nesting_depth; + goto add; + case CPP_CLOSE_PAREN: + if (nesting_depth-- == 0) + break; + goto add; + default: + add: + if (!skip) + tokens->safe_push (*token); + c_parser_consume_token (parser); + continue; + } + break; + } + + c_parser_consume_token (parser); + + if (!skip) + { + c_token eol_token; + memset (&eol_token, 0, sizeof (eol_token)); + eol_token.type = CPP_PRAGMA_EOL; + tokens->safe_push (eol_token); + } + } + else { + c_parser_error (parser, "expected % or %"); + return NULL_TREE; + } + } + c_parser_skip_to_pragma_eol (parser); + + /* Add the body tokens to the tokens for each candidate directive. */ + int nesting_depth = 0; + int bracket_depth = 0; + while (1) + { + int i; + auto_vec *tokens; + c_token *token = c_parser_peek_token (parser); + bool stop = false; + + if (begin_end_p) + { + /* Keep reading until '#pragma end metadirective' is read. */ + switch (token->type) + { + case CPP_PRAGMA: + if (token->pragma_kind == PRAGMA_OMP_END) + { + c_token *next_token = c_parser_peek_2nd_token (parser); + if (next_token->type == CPP_NAME + && strcmp (IDENTIFIER_POINTER (next_token->value), + "metadirective") == 0) + { + c_parser_consume_pragma (parser); + c_parser_consume_token (parser); + c_parser_skip_to_pragma_eol (parser); + break; + } + } + default: + FOR_EACH_VEC_ELT (directive_tokens, i, tokens) + tokens->safe_push (*token); + if (token->type == CPP_PRAGMA) + c_parser_consume_pragma (parser); + else if (token->type == CPP_PRAGMA_EOL) + c_parser_skip_to_pragma_eol (parser); + else + c_parser_consume_token (parser); + continue; + } + break; + } + else + { + switch (token->type) + { + case CPP_EOF: + break; + case CPP_OPEN_BRACE: + ++nesting_depth; + goto add2; + case CPP_CLOSE_BRACE: + if (--nesting_depth == 0) + stop = true; + goto add2; + case CPP_OPEN_PAREN: + ++bracket_depth; + goto add2; + case CPP_CLOSE_PAREN: + --bracket_depth; + goto add2; + case CPP_SEMICOLON: + if (nesting_depth == 0 && bracket_depth == 0) + stop = true; + goto add2; + default: + add2: + FOR_EACH_VEC_ELT (directive_tokens, i, tokens) + tokens->safe_push (*token); + if (token->type == CPP_PRAGMA) + c_parser_consume_pragma (parser); + else if (token->type == CPP_PRAGMA_EOL) + c_parser_skip_to_pragma_eol (parser); + else + c_parser_consume_token (parser); + if (stop) + break; + continue; + } + break; + } + } + + /* Process each candidate directive. */ + auto_vec *tokens; + int i; + FOR_EACH_VEC_ELT (directive_tokens, i, tokens) + { + /* Make sure nothing tries to read past the end of the tokens. */ + c_token eof_token; + memset (&eof_token, 0, sizeof (eof_token)); + eof_token.type = CPP_EOF; + tokens->safe_push (eof_token); + tokens->safe_push (eof_token); + + unsigned int tokens_avail = parser->tokens_avail; + gcc_assert (parser->tokens == &parser->tokens_buf[0]); + parser->tokens = tokens->address (); + parser->tokens_avail = tokens->length (); + + tree block = c_begin_compound_stmt (false); + c_parser_omp_construct (parser, if_p); + block = c_end_compound_stmt (loc, block, false); + + tree variant = build_tree_list (ctxs[i], block); + OMP_METADIRECTIVE_CLAUSES (ret) + = chainon (OMP_METADIRECTIVE_CLAUSES (ret), variant); + + parser->tokens = &parser->tokens_buf[0]; + parser->tokens_avail = tokens_avail; + } + + if (all_selectors_resolveable) + { + ret = omp_resolve_metadirective (ret); + gcc_assert (ret != NULL_TREE); + } + add_stmt (ret); + + return ret; +} + /* Main entry point to parsing most OpenMP pragmas. */ static void @@ -21676,6 +22028,11 @@ c_parser_omp_construct (c_parser *parser, bool *if_p) strcpy (p_name, "#pragma omp"); stmt = c_parser_omp_master (loc, parser, p_name, mask, NULL, if_p); break; + case PRAGMA_OMP_METADIRECTIVE: + strcpy (p_name, "#pragma omp"); + stmt = c_parser_omp_metadirective (loc, parser, p_name, mask, NULL, + if_p, false); + break; case PRAGMA_OMP_PARALLEL: strcpy (p_name, "#pragma omp"); stmt = c_parser_omp_parallel (loc, parser, p_name, mask, NULL, if_p); @@ -21713,7 +22070,6 @@ c_parser_omp_construct (c_parser *parser, bool *if_p) gcc_assert (EXPR_LOCATION (stmt) != UNKNOWN_LOCATION); } - /* OpenMP 2.5: # pragma omp threadprivate (variable-list) */ diff --git a/gcc/cp/parser.c b/gcc/cp/parser.c index f3503b13a5a..e2584c18571 100644 --- a/gcc/cp/parser.c +++ b/gcc/cp/parser.c @@ -43543,6 +43543,32 @@ cp_parser_omp_end_declare_target (cp_parser *parser, cp_token *pragma_tok) scope_chain->omp_declare_target_attribute--; } +static void +cp_parser_omp_end (cp_parser *parser, cp_token *pragma_tok) +{ + const char *p = ""; + if (cp_lexer_next_token_is (parser->lexer, CPP_NAME)) + { + tree id = cp_lexer_peek_token (parser->lexer)->u.value; + p = IDENTIFIER_POINTER (id); + } + if (strcmp (p, "declare") == 0) + { + cp_parser_omp_end_declare_target (parser, pragma_tok); + return; + } + else if (strcmp (p, "metadirective") == 0) + /* The pragma 'omp end metadirective' should have been consumed + when processing the metadirective. */ + error_at (pragma_tok->location, + "%<#pragma omp end metadirective%> without corresponding " + "%<#pragma omp begin metadirective%>"); + else + error_at (pragma_tok->location, + "expected % or %"); + cp_parser_skip_to_pragma_eol (parser, pragma_tok); +} + /* Helper function of cp_parser_omp_declare_reduction. Parse the combiner expression and optional initializer clause of #pragma omp declare reduction. We store the expression(s) as @@ -45259,8 +45285,8 @@ cp_parser_pragma (cp_parser *parser, enum pragma_context context, bool *if_p) pop_omp_privatization_clauses (stmt); return ret; - case PRAGMA_OMP_END_DECLARE_TARGET: - cp_parser_omp_end_declare_target (parser, pragma_tok); + case PRAGMA_OMP_END: + cp_parser_omp_end (parser, pragma_tok); return false; case PRAGMA_OMP_SCAN: diff --git a/gcc/gimple-low.c b/gcc/gimple-low.c index fa7d4de30c0..b68b0cbb06c 100644 --- a/gcc/gimple-low.c +++ b/gcc/gimple-low.c @@ -234,6 +234,39 @@ lower_omp_directive (gimple_stmt_iterator *gsi, struct lower_data *data) gsi_next (gsi); } +/* Lower the OpenMP metadirective statement pointed by GSI. */ + +static void +lower_omp_metadirective (gimple_stmt_iterator *gsi, struct lower_data *data) +{ + gimple *stmt = gsi_stmt (*gsi); + gimple *body = gimple_omp_metadirective_bodies (stmt); + tree succ_label = create_artificial_label (UNKNOWN_LOCATION); + unsigned i; + + for (i = 0; i < gimple_omp_metadirective_num_clauses (stmt); i++) + { + tree label = create_artificial_label (UNKNOWN_LOCATION); + gimple *g = gimple_build_label (label); + + gsi_insert_after (gsi, g, GSI_CONTINUE_LINKING); + lower_sequence (gimple_omp_body_ptr (body), data); + gsi_insert_seq_after (gsi, gimple_omp_body (body), GSI_CONTINUE_LINKING); + gsi_insert_after (gsi, gimple_build_goto (succ_label), + GSI_CONTINUE_LINKING); + gimple_omp_metadirective_set_label (stmt, i, label); + + body = body->next; + } + + gsi_insert_after (gsi, gimple_build_label (succ_label), + GSI_CONTINUE_LINKING); + gimple_omp_metadirective_set_succ_label (stmt, succ_label); + gimple_omp_metadirective_set_bodies (stmt, NULL); + + gsi_next (gsi); +} + /* Lower statement GSI. DATA is passed through the recursion. We try to track the fallthruness of statements and get rid of unreachable return @@ -398,6 +431,12 @@ lower_stmt (gimple_stmt_iterator *gsi, struct lower_data *data) data->cannot_fallthru = false; return; + case GIMPLE_OMP_METADIRECTIVE: + data->cannot_fallthru = false; + lower_omp_metadirective (gsi, data); + data->cannot_fallthru = false; + return; + case GIMPLE_TRANSACTION: lower_sequence (gimple_transaction_body_ptr ( as_a (stmt)), diff --git a/gcc/gimple-pretty-print.c b/gcc/gimple-pretty-print.c index 0ef01e6420b..e7a4ba1171b 100644 --- a/gcc/gimple-pretty-print.c +++ b/gcc/gimple-pretty-print.c @@ -1978,6 +1978,64 @@ dump_gimple_omp_return (pretty_printer *buffer, const gimple *gs, int spc, } } +/* Dump a GIMPLE_OMP_METADIRECTIVE tuple on the pretty_printer BUFFER. */ + +static void +dump_gimple_omp_metadirective (pretty_printer *buffer, const gimple *gs, + int spc, dump_flags_t flags) +{ + if (flags & TDF_RAW) + { + dump_gimple_fmt (buffer, spc, flags, "%G <%+BODY <%S> >", gs, + gimple_omp_body (gs)); + } + else + { + pp_string (buffer, "#pragma omp metadirective"); + newline_and_indent (buffer, spc + 2); + + gimple *body = gimple_omp_metadirective_bodies (gs); + bool has_bodies_p = body != NULL; + unsigned num_clauses = gimple_omp_metadirective_num_clauses (gs); + + for (unsigned i = 0; i < num_clauses; i++) + { + tree selector = gimple_omp_metadirective_selector (gs, i); + + if (selector == NULL_TREE) + pp_string (buffer, "default:"); + else + { + pp_string (buffer, "when ("); + dump_generic_node (buffer, selector, spc, flags, false); + pp_string (buffer, "):"); + } + + if (has_bodies_p) + { + newline_and_indent (buffer, spc + 4); + pp_left_brace (buffer); + pp_newline (buffer); + dump_gimple_seq (buffer, gimple_omp_body (body), spc + 6, flags); + newline_and_indent (buffer, spc + 4); + pp_right_brace (buffer); + + body = body->next; + if (body) + newline_and_indent (buffer, spc + 2); + } + else + { + tree label = gimple_omp_metadirective_label (gs, i); + pp_string (buffer, " "); + dump_generic_node (buffer, label, spc, flags, false); + if (i != num_clauses - 1) + newline_and_indent (buffer, spc + 2); + } + } + } +} + /* Dump a GIMPLE_TRANSACTION tuple on the pretty_printer BUFFER. */ static void @@ -2729,6 +2787,12 @@ pp_gimple_stmt_1 (pretty_printer *buffer, const gimple *gs, int spc, flags); break; + case GIMPLE_OMP_METADIRECTIVE: + dump_gimple_omp_metadirective (buffer, + as_a (gs), + spc, flags); + break; + case GIMPLE_CATCH: dump_gimple_catch (buffer, as_a (gs), spc, flags); break; diff --git a/gcc/gimple-streamer-in.c b/gcc/gimple-streamer-in.c index 1c979f438a5..66bbc2e8e0a 100644 --- a/gcc/gimple-streamer-in.c +++ b/gcc/gimple-streamer-in.c @@ -151,6 +151,7 @@ input_gimple_stmt (class lto_input_block *ib, class data_in *data_in, case GIMPLE_COND: case GIMPLE_GOTO: case GIMPLE_DEBUG: + case GIMPLE_OMP_METADIRECTIVE: for (i = 0; i < num_ops; i++) { tree *opp, op = stream_read_tree (ib, data_in); @@ -188,6 +189,10 @@ input_gimple_stmt (class lto_input_block *ib, class data_in *data_in, else gimple_call_set_fntype (call_stmt, stream_read_tree (ib, data_in)); } + if (gomp_metadirective *metadirective_stmt + = dyn_cast (stmt)) + gimple_omp_metadirective_set_succ_label (metadirective_stmt, + stream_read_tree (ib, data_in)); break; case GIMPLE_NOP: diff --git a/gcc/gimple-streamer-out.c b/gcc/gimple-streamer-out.c index fcbf92300d4..f0ddd6a81a3 100644 --- a/gcc/gimple-streamer-out.c +++ b/gcc/gimple-streamer-out.c @@ -127,6 +127,7 @@ output_gimple_stmt (struct output_block *ob, struct function *fn, gimple *stmt) case GIMPLE_COND: case GIMPLE_GOTO: case GIMPLE_DEBUG: + case GIMPLE_OMP_METADIRECTIVE: for (i = 0; i < gimple_num_ops (stmt); i++) { tree op = gimple_op (stmt, i); @@ -169,6 +170,8 @@ output_gimple_stmt (struct output_block *ob, struct function *fn, gimple *stmt) else stream_write_tree (ob, gimple_call_fntype (stmt), true); } + if (gimple_code (stmt) == GIMPLE_OMP_METADIRECTIVE) + stream_write_tree (ob, gimple_omp_metadirective_succ_label (stmt), true); break; case GIMPLE_NOP: diff --git a/gcc/gimple-walk.c b/gcc/gimple-walk.c index e4a55f1eeb6..ad7e1c0839e 100644 --- a/gcc/gimple-walk.c +++ b/gcc/gimple-walk.c @@ -674,6 +674,21 @@ walk_gimple_stmt (gimple_stmt_iterator *gsi, walk_stmt_fn callback_stmt, return wi->callback_result; break; + case GIMPLE_OMP_METADIRECTIVE: + { + gimple *body = gimple_omp_metadirective_bodies (stmt); + + while (body) + { + ret = walk_gimple_seq_mod (gimple_omp_body_ptr (body), + callback_stmt, callback_op, wi); + if (ret) + return wi->callback_result; + body = body->next; + } + } + break; + case GIMPLE_WITH_CLEANUP_EXPR: ret = walk_gimple_seq_mod (gimple_wce_cleanup_ptr (stmt), callback_stmt, callback_op, wi); diff --git a/gcc/gimple.c b/gcc/gimple.c index f1044e9c630..f94009d39f3 100644 --- a/gcc/gimple.c +++ b/gcc/gimple.c @@ -1234,6 +1234,28 @@ gimple_build_omp_atomic_store (tree val, enum omp_memory_order mo) return p; } +/* Build a GIMPLE_OMP_METADIRECTIVE statement. */ + +gomp_metadirective * +gimple_build_omp_metadirective (int clause_count) +{ + gomp_metadirective *p + = as_a (gimple_alloc (GIMPLE_OMP_METADIRECTIVE, + clause_count * 2)); + gimple_omp_metadirective_set_bodies (p, NULL); + return p; +} + + +gomp_metadirective_body * +gimple_build_omp_metadirective_body (gimple_seq body) +{ + gomp_metadirective_body *m_body = as_a + (gimple_alloc (GIMPLE_OMP_METADIRECTIVE_BODY, 0)); + gimple_omp_set_body (m_body, body); + return m_body; +} + /* Build a GIMPLE_TRANSACTION statement. */ gtransaction * diff --git a/gcc/gimple.def b/gcc/gimple.def index 0ac0cf72bfa..1da68c16a91 100644 --- a/gcc/gimple.def +++ b/gcc/gimple.def @@ -384,6 +384,13 @@ DEFGSCODE(GIMPLE_OMP_TEAMS, "gimple_omp_teams", GSS_OMP_PARALLEL_LAYOUT) CLAUSES is an OMP_CLAUSE chain holding the associated clauses. */ DEFGSCODE(GIMPLE_OMP_ORDERED, "gimple_omp_ordered", GSS_OMP_SINGLE_LAYOUT) +/* GIMPLE_OMP_METADIRECTIVE represents #pragma omp metadirective. */ +DEFGSCODE(GIMPLE_OMP_METADIRECTIVE, "gimple_omp_metadirective", + GSS_OMP_METADIRECTIVE) + +DEFGSCODE(GIMPLE_OMP_METADIRECTIVE_BODY, "gimple_omp_metadirective_body", + GSS_OMP_METADIRECTIVE_BODY) + /* GIMPLE_PREDICT specifies a hint for branch prediction. PREDICT is one of the predictors from predict.def. diff --git a/gcc/gimple.h b/gcc/gimple.h index 91b92b4a4d1..c5288af2bd7 100644 --- a/gcc/gimple.h +++ b/gcc/gimple.h @@ -822,6 +822,29 @@ struct GTY((tag("GSS_OMP_ATOMIC_STORE_LAYOUT"))) stmt->code == GIMPLE_OMP_RETURN. */ }; +struct GTY((tag("GSS_OMP_METADIRECTIVE_BODY"))) + gomp_metadirective_body : public gimple_statement_omp_single_layout +{ + /* No extra fields; adds invariant: + stmt->code == GIMPLE_OMP_METADIRECTIVE_BODY. */ +}; + +struct GTY((tag("GSS_OMP_METADIRECTIVE"))) + gomp_metadirective : public gimple_statement_with_ops_base +{ + /* [ WORD 1-7 ] : base class */ + + /* [ WORD 8 ] */ + gomp_metadirective_body *bodies; + + /* [ WORD 9 ] : a label after the metadirective + and all the candidate bodies */ + tree succ_label; + + /* [ WORD 10 ] : operand vector. */ + tree GTY((length ("%h.num_ops"))) op[1]; +}; + /* GIMPLE_TRANSACTION. */ /* Bits to be stored in the GIMPLE_TRANSACTION subcode. */ @@ -1233,6 +1256,22 @@ is_a_helper ::test (gimple *gs) return gs->code == GIMPLE_OMP_TASK; } +template <> +template <> +inline bool +is_a_helper ::test (gimple *gs) +{ + return gs->code == GIMPLE_OMP_METADIRECTIVE; +} + +template <> +template <> +inline bool +is_a_helper ::test (gimple *gs) +{ + return gs->code == GIMPLE_OMP_METADIRECTIVE_BODY; +} + template <> template <> inline bool @@ -1475,6 +1514,22 @@ is_a_helper ::test (const gimple *gs) return gs->code == GIMPLE_OMP_TASK; } +template <> +template <> +inline bool +is_a_helper ::test (const gimple *gs) +{ + return gs->code == GIMPLE_OMP_METADIRECTIVE; +} + +template <> +template <> +inline bool +is_a_helper ::test (const gimple *gs) +{ + return gs->code == GIMPLE_OMP_METADIRECTIVE_BODY; +} + template <> template <> inline bool @@ -1572,6 +1627,8 @@ gomp_teams *gimple_build_omp_teams (gimple_seq, 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); +gomp_metadirective *gimple_build_omp_metadirective (int clause_count); +gomp_metadirective_body *gimple_build_omp_metadirective_body (gimple_seq body); 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 *); @@ -1827,6 +1884,7 @@ gimple_has_substatements (gimple *g) case GIMPLE_OMP_TARGET: case GIMPLE_OMP_TEAMS: case GIMPLE_OMP_CRITICAL: + case GIMPLE_OMP_METADIRECTIVE: case GIMPLE_WITH_CLEANUP_EXPR: case GIMPLE_TRANSACTION: return true; @@ -2479,12 +2537,21 @@ gimple_ops (gimple *gs) } +/* Return true if GIMPLE statement G has any operands, including any that + should not be processed by the SSA passes. */ + +static inline bool +gimple_has_ops_1 (const gimple *g) +{ + return gimple_has_ops (g) || gimple_code (g) == GIMPLE_OMP_METADIRECTIVE; +} + /* Return operand I for statement GS. */ static inline tree gimple_op (const gimple *gs, unsigned i) { - if (gimple_has_ops (gs)) + if (gimple_has_ops_1 (gs)) { gcc_gimple_checking_assert (i < gimple_num_ops (gs)); return gimple_ops (CONST_CAST_GIMPLE (gs))[i]; @@ -2498,7 +2565,7 @@ gimple_op (const gimple *gs, unsigned i) static inline tree * gimple_op_ptr (gimple *gs, unsigned i) { - if (gimple_has_ops (gs)) + if (gimple_has_ops_1 (gs)) { gcc_gimple_checking_assert (i < gimple_num_ops (gs)); return gimple_ops (gs) + i; @@ -2512,7 +2579,7 @@ gimple_op_ptr (gimple *gs, unsigned i) static inline void gimple_set_op (gimple *gs, unsigned i, tree op) { - gcc_gimple_checking_assert (gimple_has_ops (gs) && i < gimple_num_ops (gs)); + gcc_gimple_checking_assert (gimple_has_ops_1 (gs) && i < gimple_num_ops (gs)); /* Note. It may be tempting to assert that OP matches is_gimple_operand, but that would be wrong. Different tuples @@ -6330,6 +6397,77 @@ gimple_omp_continue_set_control_use (gomp_continue *cont_stmt, tree use) cont_stmt->control_use = use; } + +static inline tree +gimple_omp_metadirective_succ_label (const gimple *g) +{ + const gomp_metadirective *omp_metadirective + = as_a (g); + return omp_metadirective->succ_label; +} + + +static inline void +gimple_omp_metadirective_set_succ_label (gimple *g, tree succ_label) +{ + gomp_metadirective *omp_metadirective = as_a (g); + omp_metadirective->succ_label = succ_label; +} + + +static inline gomp_metadirective_body * +gimple_omp_metadirective_bodies (const gimple *g) +{ + const gomp_metadirective *omp_metadirective + = as_a (g); + return omp_metadirective->bodies; +} + + +static inline void +gimple_omp_metadirective_set_bodies (gimple *g, + gomp_metadirective_body *bodies) +{ + gomp_metadirective *omp_metadirective = as_a (g); + omp_metadirective->bodies = bodies; +} + + +static inline unsigned +gimple_omp_metadirective_num_clauses (const gimple *g) +{ + return gimple_num_ops (g) / 2; +} + + +static inline tree +gimple_omp_metadirective_selector (const gimple *g, unsigned n) +{ + return gimple_op (g, n * 2); +} + + +static inline void +gimple_omp_metadirective_set_selector (gimple *g, unsigned n, tree selector) +{ + return gimple_set_op (g, n * 2, selector); +} + + +static inline tree +gimple_omp_metadirective_label (const gimple *g, unsigned n) +{ + return gimple_op (g, n * 2 + 1); +} + + +static inline void +gimple_omp_metadirective_set_label (gimple *g, unsigned n, tree label) +{ + return gimple_set_op (g, n * 2 + 1, label); +} + + /* Return a pointer to the body for the GIMPLE_TRANSACTION statement TRANSACTION_STMT. */ @@ -6478,6 +6616,7 @@ gimple_return_set_retval (greturn *gs, tree retval) case GIMPLE_OMP_RETURN: \ case GIMPLE_OMP_ATOMIC_LOAD: \ case GIMPLE_OMP_ATOMIC_STORE: \ + case GIMPLE_OMP_METADIRECTIVE: \ case GIMPLE_OMP_CONTINUE static inline bool diff --git a/gcc/gimplify.c b/gcc/gimplify.c index 2730f225187..49f4bcb01f3 100644 --- a/gcc/gimplify.c +++ b/gcc/gimplify.c @@ -5646,6 +5646,7 @@ is_gimple_stmt (tree t) case OMP_TASKGROUP: case OMP_ORDERED: case OMP_CRITICAL: + case OMP_METADIRECTIVE: case OMP_TASK: case OMP_TARGET: case OMP_TARGET_DATA: @@ -13785,6 +13786,49 @@ gimplify_omp_ordered (tree expr, gimple_seq body) return gimple_build_omp_ordered (body, OMP_ORDERED_CLAUSES (expr)); } +static enum gimplify_status +gimplify_omp_metadirective (tree *expr_p, gimple_seq *pre_p, gimple_seq *post_p, + bool (*gimple_test_f) (tree), fallback_t fallback) +{ + gomp_metadirective_body *first_body = NULL; + gomp_metadirective_body *prev_body = NULL; + auto_vec selectors; + unsigned i; + tree clause = OMP_METADIRECTIVE_CLAUSES (*expr_p); + + while (clause) + { + tree selector = TREE_PURPOSE (clause); + tree directive = TREE_VALUE (clause); + + selectors.safe_push (selector); + gomp_metadirective_body *body + = gimple_build_omp_metadirective_body (NULL); + gimplify_stmt (&directive, gimple_omp_body_ptr (body)); + if (!first_body) + first_body = body; + if (prev_body) + { + prev_body->next = body; + body->prev = prev_body; + } + prev_body = body; + + clause = TREE_CHAIN (clause); + } + + gomp_metadirective *stmt + = gimple_build_omp_metadirective (selectors.length ()); + gimple_omp_metadirective_set_bodies (stmt, first_body); + gimplify_seq_add_stmt (pre_p, stmt); + + tree selector; + FOR_EACH_VEC_ELT (selectors, i, selector) + gimple_omp_metadirective_set_selector (stmt, i, selector); + + return GS_ALL_DONE; +} + /* Convert the GENERIC expression tree *EXPR_P to GIMPLE. If the expression produces a value to be used as an operand inside a GIMPLE statement, the value will be stored back in *EXPR_P. This value will @@ -14680,6 +14724,11 @@ gimplify_expr (tree *expr_p, gimple_seq *pre_p, gimple_seq *post_p, ret = gimplify_omp_atomic (expr_p, pre_p); break; + case OMP_METADIRECTIVE: + ret = gimplify_omp_metadirective (expr_p, pre_p, post_p, + gimple_test_f, fallback); + break; + case TRANSACTION_EXPR: ret = gimplify_transaction (expr_p, pre_p); break; diff --git a/gcc/gsstruct.def b/gcc/gsstruct.def index 8f777e2bb95..f22ac1f65cb 100644 --- a/gcc/gsstruct.def +++ b/gcc/gsstruct.def @@ -50,4 +50,6 @@ DEFGSSTRUCT(GSS_OMP_SINGLE_LAYOUT, gimple_statement_omp_single_layout, false) DEFGSSTRUCT(GSS_OMP_CONTINUE, gomp_continue, false) DEFGSSTRUCT(GSS_OMP_ATOMIC_LOAD, gomp_atomic_load, false) DEFGSSTRUCT(GSS_OMP_ATOMIC_STORE_LAYOUT, gomp_atomic_store, false) +DEFGSSTRUCT(GSS_OMP_METADIRECTIVE, gomp_metadirective, true) +DEFGSSTRUCT(GSS_OMP_METADIRECTIVE_BODY, gomp_metadirective_body, false) DEFGSSTRUCT(GSS_TRANSACTION, gtransaction, false) diff --git a/gcc/omp-expand-metadirective.cc b/gcc/omp-expand-metadirective.cc new file mode 100644 index 00000000000..f4620df577a --- /dev/null +++ b/gcc/omp-expand-metadirective.cc @@ -0,0 +1,125 @@ +/* Expand an OpenMP metadirective. + + Copyright (C) 2021 Free Software Foundation, Inc. + +This file is part of GCC. + +GCC is free software; you can redistribute it and/or modify it under +the terms of the GNU General Public License as published by the Free +Software Foundation; either version 3, or (at your option) any later +version. + +GCC is distributed in the hope that it will be useful, but WITHOUT ANY +WARRANTY; without even the implied warranty of MERCHANTABILITY or +FITNESS FOR A PARTICULAR PURPOSE. See the GNU General Public License +for more details. + +You should have received a copy of the GNU General Public License +along with GCC; see the file COPYING3. If not see +. */ + +#include "config.h" +#include "system.h" +#include "coretypes.h" +#include "backend.h" +#include "target.h" +#include "tree.h" +#include "langhooks.h" +#include "gimple.h" +#include "tree-pass.h" +#include "cgraph.h" +#include "fold-const.h" +#include "gimplify.h" +#include "gimple-iterator.h" +#include "gimple-walk.h" +#include "gomp-constants.h" +#include "omp-general.h" +#include "diagnostic-core.h" +#include "tree-cfg.h" +#include "cfganal.h" + +static void +omp_expand_metadirective (function *fun, basic_block bb) +{ + gimple *stmt = last_stmt (bb); + tree selected_label = omp_resolve_metadirective (stmt); + + /* This is the last chance for the metadirective to be resolved. */ + if (!selected_label) + gcc_unreachable (); + + /* Delete all variant BBs except for the selected one. */ + calculate_dominance_info (CDI_DOMINATORS); + for (unsigned i = 0; i < gimple_omp_metadirective_num_clauses (stmt); i++) + { + tree label = gimple_omp_metadirective_label (stmt, i); + edge edge = find_edge (bb, label_to_block (fun, label)); + if (label == selected_label) + edge->flags |= EDGE_FALLTHRU; + else + remove_edge_and_dominated_blocks (edge); + } + + /* Remove the metadirective statement. */ + gimple_stmt_iterator gsi = gsi_last_bb (bb); + gsi_remove (&gsi, true); +} + +namespace { + +const pass_data pass_data_omp_expand_metadirective = +{ + GIMPLE_PASS, /* type */ + "omp_expand_metadirective", /* name */ + OPTGROUP_OMP, /* optinfo_flags */ + TV_NONE, /* tv_id */ + PROP_gimple_lcf, /* properties_required */ + 0, /* properties_provided */ + 0, /* properties_destroyed */ + 0, /* todo_flags_start */ + 0, /* todo_flags_finish */ +}; + +class pass_omp_expand_metadirective : public gimple_opt_pass +{ +public: + pass_omp_expand_metadirective (gcc::context *ctxt) + : gimple_opt_pass (pass_data_omp_expand_metadirective, ctxt) + {} + + /* opt_pass methods: */ + virtual bool gate (function *) + { + return (flag_openmp); + } + + virtual unsigned int execute (function *fun); +}; // class pass_omp_oacc_kernels_decompose + +unsigned int +pass_omp_expand_metadirective::execute (function *fun) +{ + basic_block bb; + auto_vec metadirective_bbs; + + FOR_EACH_BB_FN (bb, fun) + { + gimple *stmt = last_stmt (bb); + if (stmt && is_a (stmt)) + metadirective_bbs.safe_push (bb); + } + + for (unsigned i = 0; i < metadirective_bbs.length (); i++) + omp_expand_metadirective (fun, metadirective_bbs[i]); + + return 0; +} + +} // anon namespace + + +gimple_opt_pass * +make_pass_omp_expand_metadirective (gcc::context *ctxt) +{ + return new pass_omp_expand_metadirective (ctxt); +} diff --git a/gcc/omp-expand.c b/gcc/omp-expand.c index 0f843bad79a..2c1affb64f8 100644 --- a/gcc/omp-expand.c +++ b/gcc/omp-expand.c @@ -9866,6 +9866,22 @@ expand_omp_target (struct omp_region *region) } } +static void +expand_omp_metadirective (struct omp_region *region) +{ + gomp_metadirective *stmt + = as_a (last_stmt (region->entry)); + tree succ_label = gimple_omp_metadirective_succ_label (stmt); + basic_block succ_bb = label_to_block (cfun, succ_label); + gimple_stmt_iterator gsi = gsi_start_bb (succ_bb); + while (!gsi_end_p (gsi) + && gimple_code (gsi_stmt (gsi)) != GIMPLE_OMP_RETURN) + gsi_next (&gsi); + + gcc_assert (gimple_code (gsi_stmt (gsi)) == GIMPLE_OMP_RETURN); + gsi_remove (&gsi, true); +} + /* Expand the parallel region tree rooted at REGION. Expansion proceeds in depth-first order. Innermost regions are expanded first. This way, parallel regions that require a new function to @@ -9952,6 +9968,10 @@ expand_omp (struct omp_region *region) expand_omp_target (region); break; + case GIMPLE_OMP_METADIRECTIVE: + expand_omp_metadirective (region); + break; + default: gcc_unreachable (); } @@ -10388,6 +10408,24 @@ omp_make_gimple_edges (basic_block bb, struct omp_region **region, } break; + case GIMPLE_OMP_METADIRECTIVE: + /* Create an edge to the beginning of the body of each candidate + directive. */ + { + unsigned i; + cur_region = new_omp_region (bb, code, cur_region); + gimple *stmt = last_stmt (bb); + for (i = 0; i < gimple_omp_metadirective_num_clauses (stmt); i++) + { + tree dest = gimple_omp_metadirective_label (stmt, i); + basic_block dest_bb = label_to_block (cfun, dest); + make_edge (bb, dest_bb, 0); + } + + fallthru = false; + } + break; + default: gcc_unreachable (); } diff --git a/gcc/omp-general.c b/gcc/omp-general.c index a1bb9d8d25d..4839a9849bc 100644 --- a/gcc/omp-general.c +++ b/gcc/omp-general.c @@ -44,6 +44,7 @@ along with GCC; see the file COPYING3. If not see #include "tree-iterator.h" #include "data-streamer.h" #include "streamer-hooks.h" +#include "tree-pretty-print.h" enum omp_requires omp_requires_mask; @@ -1100,8 +1101,13 @@ omp_context_name_list_prop (tree prop) others need to wait until the whole TU is parsed, others need to wait until IPA, others until vectorization. */ +#define DELAY_METADIRECTIVES_AFTER_LTO { \ + if (metadirective_p && !(cfun->curr_properties & PROP_gimple_lomp_dev)) \ + return -1; \ +} + int -omp_context_selector_matches (tree ctx) +omp_context_selector_matches (tree ctx, bool metadirective_p) { int ret = 1; for (tree t1 = ctx; t1; t1 = TREE_CHAIN (t1)) @@ -1222,6 +1228,8 @@ omp_context_selector_matches (tree ctx) const char *arch = omp_context_name_list_prop (t3); if (arch == NULL) return 0; + DELAY_METADIRECTIVES_AFTER_LTO; + int r = 0; if (targetm.omp.device_kind_arch_isa != NULL) r = targetm.omp.device_kind_arch_isa (omp_device_arch, @@ -1340,6 +1348,8 @@ omp_context_selector_matches (tree ctx) return 0; continue; } + DELAY_METADIRECTIVES_AFTER_LTO; + int r = 0; if (targetm.omp.device_kind_arch_isa != NULL) r = targetm.omp.device_kind_arch_isa (omp_device_kind, @@ -1379,6 +1389,8 @@ omp_context_selector_matches (tree ctx) const char *isa = omp_context_name_list_prop (t3); if (isa == NULL) return 0; + DELAY_METADIRECTIVES_AFTER_LTO; + int r = 0; if (targetm.omp.device_kind_arch_isa != NULL) r = targetm.omp.device_kind_arch_isa (omp_device_isa, @@ -1445,6 +1457,8 @@ omp_context_selector_matches (tree ctx) return ret; } +#undef DELAY_METADIRECTIVES_AFTER_LTO + /* Compare construct={simd} CLAUSES1 with CLAUSES2, return 0/-1/1/2 as in omp_context_selector_set_compare. */ @@ -2459,6 +2473,161 @@ omp_lto_input_declare_variant_alt (lto_input_block *ib, cgraph_node *node, INSERT) = entryp; } +tree +omp_resolve_metadirective (tree metadirective) +{ + auto_vec clauses; + auto_vec scores; + tree clause = OMP_METADIRECTIVE_CLAUSES (metadirective); + tree default_variant = NULL_TREE; + + while (clause) + { + tree selector = TREE_PURPOSE (clause); + widest_int score; + + if (selector == NULL_TREE) + default_variant = TREE_VALUE (clause); + else + switch (omp_context_selector_matches (selector, true)) + { + case -1: + return NULL_TREE; + case 1: + clauses.safe_push (clause); + /* TODO: Handle SIMD score? */ + omp_context_compute_score (selector, &score, false); + scores.safe_push (score); + break; + case 0: + break; + } + clause = TREE_CHAIN (clause); + } + + /* TODO: Handle case where there is no default. */ + if (clauses.is_empty ()) + { + if (dump_file) + fprintf (dump_file, "Selecting default directive variant\n"); + return default_variant; + } + + /* A context selector that is a strict subset of another context selector + has a score of zero. */ + tree clause1, clause2; + unsigned int i, j; + FOR_EACH_VEC_ELT (clauses, i, clause1) + FOR_EACH_VEC_ELT_FROM (clauses, j, clause2, i + 1) + { + int r = omp_context_selector_compare (TREE_PURPOSE (clause1), + TREE_PURPOSE (clause2)); + if (r == -1) + { + /* ctx1 is a strict subset of ctx2. */ + scores[i] = 0; + break; + } + else if (r == 1) + /* ctx2 is a strict subset of ctx1. */ + scores[j] = 0; + } + + widest_int score, highest_score = -1; + FOR_EACH_VEC_ELT (scores, i, score) + if (score > highest_score) + { + highest_score = score; + clause = clauses[i]; + } + + if (dump_file) + { + fprintf (dump_file, "Selecting directive variant with selector:"); + print_generic_expr (dump_file, TREE_PURPOSE (clause)); + fprintf (dump_file, "\n"); + } + return TREE_VALUE (clause); +} + +tree +omp_resolve_metadirective (gimple *gs) +{ + auto_vec labels; + auto_vec selectors; + auto_vec scores; + tree default_label = gimple_omp_metadirective_succ_label (gs); + + for (unsigned i = 0; i < gimple_omp_metadirective_num_clauses (gs); i++) + { + tree selector = gimple_omp_metadirective_selector (gs, i); + widest_int score; + if (selector == NULL_TREE) + default_label = gimple_omp_metadirective_label (gs, i); + else + switch (omp_context_selector_matches (selector, true)) + { + case -1: + return NULL; + case 1: + labels.safe_push (gimple_omp_metadirective_label (gs, i)); + selectors.safe_push (selector); + /* TODO: Handle SIMD score?. */ + omp_context_compute_score (selector, &score, false); + scores.safe_push (score); + break; + case 0: + break; + } + } + + if (scores.is_empty ()) + { + if (dump_file) + fprintf (dump_file, "Selecting default directive variant\n"); + return default_label; + } + + /* A context selector that is a strict subset of another context selector + has a score of zero. */ + tree ctx1, ctx2; + unsigned int i, j; + FOR_EACH_VEC_ELT (selectors, i, ctx1) + FOR_EACH_VEC_ELT_FROM (selectors, j, ctx2, i + 1) + { + int r = omp_context_selector_compare (ctx1, ctx2); + if (r == -1) + { + /* ctx1 is a strict subset of ctx2. */ + scores[i] = 0; + break; + } + else if (r == 1) + /* ctx2 is a strict subset of ctx1. */ + scores[j] = 0; + } + + unsigned highest_index = 0; + widest_int score, highest_score = -1; + FOR_EACH_VEC_ELT (scores, i, score) + { + if (score > highest_score) + { + highest_score = score; + highest_index = i; + } + } + + if (dump_file) + { + fprintf (dump_file, "Selecting directive variant with selector:"); + print_generic_expr (dump_file, selectors[highest_index]); + fprintf (dump_file, "\n"); + } + + return labels[highest_index]; +} + /* Encode an oacc launch argument. This matches the GOMP_LAUNCH_PACK macro on gomp-constants.h. We do not check for overflow. */ diff --git a/gcc/omp-general.h b/gcc/omp-general.h index aa04895e16d..47cea2eae01 100644 --- a/gcc/omp-general.h +++ b/gcc/omp-general.h @@ -104,10 +104,12 @@ extern tree find_combined_omp_for (tree *, int *, void *); extern poly_uint64 omp_max_vf (void); extern int omp_max_simt_vf (void); extern int omp_constructor_traits_to_codes (tree, enum tree_code *); -extern int omp_context_selector_matches (tree); +extern int omp_context_selector_matches (tree, bool = false); extern int omp_context_selector_set_compare (const char *, tree, tree); extern tree omp_get_context_selector (tree, const char *, const char *); extern tree omp_resolve_declare_variant (tree); +extern tree omp_resolve_metadirective (tree); +extern tree omp_resolve_metadirective (gimple *); extern tree oacc_launch_pack (unsigned code, tree device, unsigned op); extern tree oacc_replace_fn_attrib_attr (tree attribs, tree dims); extern void oacc_replace_fn_attrib (tree fn, tree dims); diff --git a/gcc/omp-low.c b/gcc/omp-low.c index d1136d181b3..c54000cfb9d 100644 --- a/gcc/omp-low.c +++ b/gcc/omp-low.c @@ -179,6 +179,10 @@ struct omp_context /* Only used for omp target contexts. True if an OpenMP construct other than teams is strictly nested in it. */ bool nonteams_nested_p; + + /* Only used for omp metadirectives. Links to the next shallow + clone of this context. */ + struct omp_context *next_clone; }; static splay_tree all_contexts; @@ -964,6 +968,7 @@ new_omp_context (gimple *stmt, omp_context *outer_ctx) splay_tree_insert (all_contexts, (splay_tree_key) stmt, (splay_tree_value) ctx); ctx->stmt = stmt; + ctx->next_clone = NULL; if (outer_ctx) { @@ -993,6 +998,17 @@ new_omp_context (gimple *stmt, omp_context *outer_ctx) return ctx; } +static omp_context * +clone_omp_context (omp_context *ctx) +{ + omp_context *clone_ctx = XCNEW (omp_context); + + memcpy (clone_ctx, ctx, sizeof (omp_context)); + ctx->next_clone = clone_ctx; + + return clone_ctx; +} + static gimple_seq maybe_catch_exception (gimple_seq); /* Finalize task copyfn. */ @@ -1039,6 +1055,14 @@ delete_omp_context (splay_tree_value value) { omp_context *ctx = (omp_context *) value; + /* Delete clones. */ + omp_context *clone = ctx->next_clone; + while (clone) + { + clone = clone->next_clone; + XDELETE (clone); + } + delete ctx->cb.decl_map; if (ctx->field_map) @@ -1073,6 +1097,7 @@ delete_omp_context (splay_tree_value value) delete ctx->lastprivate_conditional_map; delete ctx->allocate_map; + XDELETE (ctx); } @@ -3008,6 +3033,23 @@ scan_omp_teams (gomp_teams *stmt, omp_context *outer_ctx) ctx->record_type = ctx->receiver_decl = NULL; } +/* Scan an OpenMP metadirective. */ + +static void +scan_omp_metadirective (gomp_metadirective *stmt, omp_context *outer_ctx) +{ + gomp_metadirective_body *body = gimple_omp_metadirective_bodies (stmt); + + while (body) + { + gimple_seq *body_p = gimple_omp_body_ptr (body); + omp_context *ctx = outer_ctx ? clone_omp_context (outer_ctx) : NULL; + scan_omp (body_p, ctx); + + body = (gomp_metadirective_body *) body->next; + } +} + /* Check nesting restrictions. */ static bool check_omp_nesting_restrictions (gimple *stmt, omp_context *ctx) @@ -4045,6 +4087,10 @@ scan_omp_1_stmt (gimple_stmt_iterator *gsi, bool *handled_ops_p, scan_omp_teams (as_a (stmt), ctx); break; + case GIMPLE_OMP_METADIRECTIVE: + scan_omp_metadirective (as_a (stmt), ctx); + break; + case GIMPLE_BIND: { tree var; @@ -10129,6 +10175,22 @@ lower_omp_for_lastprivate (struct omp_for_data *fd, gimple_seq *body_p, } } +static void +lower_omp_metadirective (gimple_stmt_iterator *gsi_p, omp_context *ctx) +{ + gimple *stmt = gsi_stmt (*gsi_p); + gomp_metadirective_body *body = gimple_omp_metadirective_bodies (stmt); + while (body) + { + gimple_seq *body_p = gimple_omp_body_ptr (body); + omp_context *ctx = maybe_lookup_ctx (*body_p); + lower_omp (body_p, ctx); + body = (gomp_metadirective_body *) (body->next); + } + gsi_insert_after (gsi_p, gimple_build_omp_return (true), + GSI_CONTINUE_LINKING); +} + /* Callback for walk_gimple_seq. Find #pragma omp scan statement. */ static tree @@ -13474,6 +13536,9 @@ lower_omp_1 (gimple_stmt_iterator *gsi_p, omp_context *ctx) else lower_omp_teams (gsi_p, ctx); break; + case GIMPLE_OMP_METADIRECTIVE: + lower_omp_metadirective (gsi_p, ctx); + break; case GIMPLE_CALL: tree fndecl; call_stmt = as_a (stmt); diff --git a/gcc/passes.def b/gcc/passes.def index 945d2bc797c..9aad498f266 100644 --- a/gcc/passes.def +++ b/gcc/passes.def @@ -186,6 +186,7 @@ along with GCC; see the file COPYING3. If not see NEXT_PASS (pass_oacc_device_lower); NEXT_PASS (pass_omp_device_lower); NEXT_PASS (pass_omp_target_link); + NEXT_PASS (pass_omp_expand_metadirective); NEXT_PASS (pass_adjust_alignment); NEXT_PASS (pass_all_optimizations); PUSH_INSERT_PASSES_WITHIN (pass_all_optimizations) diff --git a/gcc/tree-cfg.c b/gcc/tree-cfg.c index 02256580c98..525e945a2d7 100644 --- a/gcc/tree-cfg.c +++ b/gcc/tree-cfg.c @@ -1668,6 +1668,23 @@ cleanup_dead_labels (void) } break; + case GIMPLE_OMP_METADIRECTIVE: + { + int i; + for (i = 0; i < gimple_omp_metadirective_num_clauses (stmt); i++) + { + label = gimple_omp_metadirective_label (stmt, i); + new_label = main_block_label (label, label_for_bb); + if (new_label != label) + gimple_omp_metadirective_set_label (stmt, i, new_label); + } + label = gimple_omp_metadirective_succ_label (stmt); + new_label = main_block_label (label, label_for_bb); + if (new_label != label) + gimple_omp_metadirective_set_succ_label (stmt, new_label); + } + break; + default: break; } @@ -6078,6 +6095,22 @@ gimple_redirect_edge_and_branch (edge e, basic_block dest) gimple_block_label (dest)); break; + case GIMPLE_OMP_METADIRECTIVE: + { + for (unsigned i = 0; i < gimple_omp_metadirective_num_clauses (stmt); i++) + { + tree label = gimple_omp_metadirective_label (stmt, i); + if (label_to_block (cfun, label) == e->dest) + gimple_omp_metadirective_set_label (stmt, i, + gimple_block_label (dest)); + } + tree label = gimple_omp_metadirective_succ_label (stmt); + if (label_to_block (cfun, label) == e->dest) + gimple_omp_metadirective_set_succ_label (stmt, + gimple_block_label (dest)); + } + break; + default: /* Otherwise it must be a fallthru edge, and we don't need to do anything besides redirecting it. */ diff --git a/gcc/tree-inline.c b/gcc/tree-inline.c index 8f945b88c12..3a21268b93b 100644 --- a/gcc/tree-inline.c +++ b/gcc/tree-inline.c @@ -4534,6 +4534,12 @@ estimate_num_insns (gimple *stmt, eni_weights *weights) return (weights->omp_cost + estimate_num_insns_seq (gimple_omp_body (stmt), weights)); + case GIMPLE_OMP_METADIRECTIVE: + /* The actual instruction will disappear eventually, so metadirective + statements have zero cost. */ + gcc_assert (gimple_omp_body (stmt) == NULL); + return 0; + case GIMPLE_TRANSACTION: return (weights->tm_cost + estimate_num_insns_seq (gimple_transaction_body ( diff --git a/gcc/tree-pass.h b/gcc/tree-pass.h index 15693fee150..c02dda89f6a 100644 --- a/gcc/tree-pass.h +++ b/gcc/tree-pass.h @@ -418,6 +418,7 @@ extern gimple_opt_pass *make_pass_lower_switch_O0 (gcc::context *ctxt); extern gimple_opt_pass *make_pass_lower_vector (gcc::context *ctxt); extern gimple_opt_pass *make_pass_lower_vector_ssa (gcc::context *ctxt); extern gimple_opt_pass *make_pass_omp_oacc_kernels_decompose (gcc::context *ctxt); +extern gimple_opt_pass *make_pass_omp_expand_metadirective (gcc::context *ctxt); extern gimple_opt_pass *make_pass_lower_omp (gcc::context *ctxt); extern gimple_opt_pass *make_pass_diagnose_omp_blocks (gcc::context *ctxt); extern gimple_opt_pass *make_pass_expand_omp (gcc::context *ctxt); diff --git a/gcc/tree-pretty-print.c b/gcc/tree-pretty-print.c index 0a575eb9dad..5b601a303c6 100644 --- a/gcc/tree-pretty-print.c +++ b/gcc/tree-pretty-print.c @@ -3626,6 +3626,34 @@ dump_generic_node (pretty_printer *pp, tree node, int spc, dump_flags_t flags, is_expr = false; break; + case OMP_METADIRECTIVE: + { + pp_string (pp, "#pragma omp metadirective"); + newline_and_indent (pp, spc + 2); + pp_left_brace (pp); + + tree clause = OMP_METADIRECTIVE_CLAUSES (node); + while (clause != NULL_TREE) + { + newline_and_indent (pp, spc + 4); + if (TREE_PURPOSE (clause) == NULL_TREE) + pp_string (pp, "default:"); + else + { + pp_string (pp, "when ("); + dump_generic_node (pp, TREE_PURPOSE (clause), spc + 4, flags, + false); + pp_string (pp, "):"); + } + newline_and_indent (pp, spc + 6); + dump_generic_node (pp, TREE_VALUE (clause), spc + 6, flags, false); + clause = TREE_CHAIN (clause); + } + newline_and_indent (pp, spc + 2); + pp_right_brace (pp); + } + break; + case TRANSACTION_EXPR: if (TRANSACTION_EXPR_OUTER (node)) pp_string (pp, "__transaction_atomic [[outer]]"); diff --git a/gcc/tree-ssa-operands.c b/gcc/tree-ssa-operands.c index c15575416dd..ee26451d717 100644 --- a/gcc/tree-ssa-operands.c +++ b/gcc/tree-ssa-operands.c @@ -978,6 +978,9 @@ operands_scanner::parse_ssa_operands () append_vuse (gimple_vop (fn)); goto do_default; + case GIMPLE_OMP_METADIRECTIVE: + break; + case GIMPLE_CALL: /* Add call-clobbered operands, if needed. */ maybe_add_call_vops (as_a (stmt)); diff --git a/gcc/tree.def b/gcc/tree.def index eda050bdc55..9d50c739539 100644 --- a/gcc/tree.def +++ b/gcc/tree.def @@ -1264,6 +1264,11 @@ DEFTREECODE (OMP_TARGET_ENTER_DATA, "omp_target_enter_data", tcc_statement, 1) Operand 0: OMP_TARGET_EXIT_DATA_CLAUSES: List of clauses. */ DEFTREECODE (OMP_TARGET_EXIT_DATA, "omp_target_exit_data", tcc_statement, 1) +/* OpenMP - #pragma omp metadirective [clause1 ... clauseN] + Operand 0: OMP_METADIRECTIVE_CLAUSES: List of selectors and directive + variants. */ +DEFTREECODE (OMP_METADIRECTIVE, "omp_metadirective", tcc_statement, 1) + /* OMP_ATOMIC through OMP_ATOMIC_CAPTURE_NEW must be consecutive, or OMP_ATOMIC_SEQ_CST needs adjusting. */ diff --git a/gcc/tree.h b/gcc/tree.h index 64612cfa368..0d74cc75ce1 100644 --- a/gcc/tree.h +++ b/gcc/tree.h @@ -1459,6 +1459,9 @@ class auto_suppress_location_wrappers #define OMP_TARGET_EXIT_DATA_CLAUSES(NODE)\ TREE_OPERAND (OMP_TARGET_EXIT_DATA_CHECK (NODE), 0) +#define OMP_METADIRECTIVE_CLAUSES(NODE) \ + TREE_OPERAND (OMP_METADIRECTIVE_CHECK (NODE), 0) + #define OMP_SCAN_BODY(NODE) TREE_OPERAND (OMP_SCAN_CHECK (NODE), 0) #define OMP_SCAN_CLAUSES(NODE) TREE_OPERAND (OMP_SCAN_CHECK (NODE), 1)