From ae9281fc64aebd40fbe0ce1705c6ca0aaffdb0c4 Mon Sep 17 00:00:00 2001 From: Thomas Schwinge Date: Thu, 4 Aug 2016 15:35:30 +0200 Subject: [PATCH] Rework C/C++ OpenACC routine parsing gcc/c/ * c-parser.c (struct oacc_routine_data): Add error_seen and fndecl_seen members. (c_finish_oacc_routine): Use these. (c_parser_declaration_or_fndef): Adjust. (c_parser_oacc_routine): Likewise. Support more C language constructs, and improve diagnostics. Move pragma context checking... (c_parser_pragma): ... here. gcc/cp/ * parser.c (cp_ensure_no_oacc_routine): Improve diagnostics. (cp_parser_late_parsing_cilk_simd_fn_info): Fix diagnostics. (cp_parser_late_parsing_oacc_routine, cp_finalize_oacc_routine): Simplify code, and improve diagnostics. (cp_parser_oacc_routine): Likewise. Move pragma context checking... (cp_parser_pragma): ... here. gcc/testsuite/ * c-c++-common/goacc/routine-5.c: Update. From-SVN: r239128 --- gcc/c/ChangeLog | 9 + gcc/c/c-parser.c | 165 ++++++++++----- gcc/cp/ChangeLog | 8 + gcc/cp/parser.c | 180 ++++++++--------- gcc/testsuite/ChangeLog | 4 + gcc/testsuite/c-c++-common/goacc/routine-5.c | 199 ++++++++++++++++--- 6 files changed, 391 insertions(+), 174 deletions(-) diff --git a/gcc/c/ChangeLog b/gcc/c/ChangeLog index 7ef094a9cf5..ecae4f158bb 100644 --- a/gcc/c/ChangeLog +++ b/gcc/c/ChangeLog @@ -1,5 +1,14 @@ 2016-08-04 Thomas Schwinge + * c-parser.c (struct oacc_routine_data): Add error_seen and + fndecl_seen members. + (c_finish_oacc_routine): Use these. + (c_parser_declaration_or_fndef): Adjust. + (c_parser_oacc_routine): Likewise. Support more C language + constructs, and improve diagnostics. Move pragma context + checking... + (c_parser_pragma): ... here. + * c-parser.c (struct oacc_routine_data): New. (c_parser_declaration_or_fndef, c_parser_oacc_routine): Use it. Simplify code. diff --git a/gcc/c/c-parser.c b/gcc/c/c-parser.c index c4a9797d49c..ec74e0b3ead 100644 --- a/gcc/c/c-parser.c +++ b/gcc/c/c-parser.c @@ -1276,6 +1276,8 @@ enum c_parser_prec { /* Helper data structure for parsing #pragma acc routine. */ struct oacc_routine_data { + bool error_seen; /* Set if error has been reported. */ + bool fndecl_seen; /* Set if one fn decl/definition has been seen already. */ tree clauses; location_t loc; }; @@ -1568,8 +1570,7 @@ c_parser_external_declaration (c_parser *parser) } static void c_finish_omp_declare_simd (c_parser *, tree, tree, vec); -static void c_finish_oacc_routine (struct oacc_routine_data *, tree, bool, - bool, bool); +static void c_finish_oacc_routine (struct oacc_routine_data *, tree, bool); /* Parse a declaration or function definition (C90 6.5, 6.7.1, C99 6.7, 6.9.1). If FNDEF_OK is true, a function definition is @@ -1754,8 +1755,7 @@ c_parser_declaration_or_fndef (c_parser *parser, bool fndef_ok, } c_parser_consume_token (parser); if (oacc_routine_data) - c_finish_oacc_routine (oacc_routine_data, NULL_TREE, false, true, - false); + c_finish_oacc_routine (oacc_routine_data, NULL_TREE, false); return; } @@ -1853,7 +1853,7 @@ c_parser_declaration_or_fndef (c_parser *parser, bool fndef_ok, prefix_attrs = specs->attrs; all_prefix_attrs = prefix_attrs; specs->attrs = NULL_TREE; - for (bool first = true;; first = false) + while (true) { struct c_declarator *declarator; bool dummy = false; @@ -1873,8 +1873,7 @@ c_parser_declaration_or_fndef (c_parser *parser, bool fndef_ok, c_finish_omp_declare_simd (parser, NULL_TREE, NULL_TREE, omp_declare_simd_clauses); if (oacc_routine_data) - c_finish_oacc_routine (oacc_routine_data, NULL_TREE, - false, first, false); + c_finish_oacc_routine (oacc_routine_data, NULL_TREE, false); c_parser_skip_to_end_of_block_or_statement (parser); return; } @@ -1990,8 +1989,7 @@ c_parser_declaration_or_fndef (c_parser *parser, bool fndef_ok, finish_init (); } if (oacc_routine_data) - c_finish_oacc_routine (oacc_routine_data, d, - false, first, false); + c_finish_oacc_routine (oacc_routine_data, d, false); if (d != error_mark_node) { maybe_warn_string_init (init_loc, TREE_TYPE (d), init); @@ -2036,8 +2034,7 @@ c_parser_declaration_or_fndef (c_parser *parser, bool fndef_ok, temp_pop_parm_decls (); } if (oacc_routine_data) - c_finish_oacc_routine (oacc_routine_data, d, - false, first, false); + c_finish_oacc_routine (oacc_routine_data, d, false); if (d) finish_decl (d, UNKNOWN_LOCATION, NULL_TREE, NULL_TREE, asm_name); @@ -2149,8 +2146,7 @@ c_parser_declaration_or_fndef (c_parser *parser, bool fndef_ok, c_finish_omp_declare_simd (parser, current_function_decl, NULL_TREE, omp_declare_simd_clauses); if (oacc_routine_data) - c_finish_oacc_routine (oacc_routine_data, current_function_decl, - false, first, true); + c_finish_oacc_routine (oacc_routine_data, current_function_decl, true); DECL_STRUCT_FUNCTION (current_function_decl)->function_start_locus = c_parser_peek_token (parser)->location; fnbody = c_parser_compound_statement (parser); @@ -10123,6 +10119,13 @@ c_parser_pragma (c_parser *parser, enum pragma_context context, bool *if_p) return false; case PRAGMA_OACC_ROUTINE: + if (context != pragma_external) + { + error_at (c_parser_peek_token (parser)->location, + "%<#pragma acc routine%> must be at file scope"); + c_parser_skip_until_found (parser, CPP_PRAGMA_EOL, NULL); + return false; + } c_parser_oacc_routine (parser, context); return false; @@ -14030,29 +14033,32 @@ c_parser_oacc_kernels_parallel (location_t loc, c_parser *parser, static void c_parser_oacc_routine (c_parser *parser, enum pragma_context context) { - tree decl = NULL_TREE; + gcc_checking_assert (context == pragma_external); + oacc_routine_data data; + data.error_seen = false; + data.fndecl_seen = false; data.clauses = NULL_TREE; data.loc = c_parser_peek_token (parser)->location; - - if (context != pragma_external) - c_parser_error (parser, "%<#pragma acc routine%> not at file scope"); c_parser_consume_pragma (parser); - /* Scan for optional '( name )'. */ - if (c_parser_peek_token (parser)->type == CPP_OPEN_PAREN) + /* Look for optional '( name )'. */ + if (c_parser_next_token_is (parser, CPP_OPEN_PAREN)) { - c_parser_consume_token (parser); + c_parser_consume_token (parser); /* '(' */ - c_token *token = c_parser_peek_token (parser); - if (token->type == CPP_NAME && (token->id_kind == C_ID_ID - || token->id_kind == C_ID_TYPENAME)) + tree decl = NULL_TREE; + c_token *name_token = c_parser_peek_token (parser); + location_t name_loc = name_token->location; + if (name_token->type == CPP_NAME + && (name_token->id_kind == C_ID_ID + || name_token->id_kind == C_ID_TYPENAME)) { - decl = lookup_name (token->value); + decl = lookup_name (name_token->value); if (!decl) - error_at (token->location, "%qE has not been declared", - token->value); + error_at (name_loc, + "%qE has not been declared", name_token->value); c_parser_consume_token (parser); } else @@ -14064,22 +14070,56 @@ c_parser_oacc_routine (c_parser *parser, enum pragma_context context) c_parser_skip_to_pragma_eol (parser, false); return; } + + data.clauses + = c_parser_oacc_all_clauses (parser, OACC_ROUTINE_CLAUSE_MASK, + "#pragma acc routine"); + + if (TREE_CODE (decl) != FUNCTION_DECL) + { + error_at (name_loc, "%qD does not refer to a function", decl); + return; + } + + c_finish_oacc_routine (&data, decl, false); } + else /* No optional '( name )'. */ + { + data.clauses + = c_parser_oacc_all_clauses (parser, OACC_ROUTINE_CLAUSE_MASK, + "#pragma acc routine"); - /* Build a chain of clauses. */ - parser->in_pragma = true; - data.clauses - = c_parser_oacc_all_clauses (parser, OACC_ROUTINE_CLAUSE_MASK, - "#pragma acc routine"); - - if (decl) - c_finish_oacc_routine (&data, decl, true, true, false); - else if (c_parser_peek_token (parser)->type == CPP_PRAGMA) - /* This will emit an error. */ - c_finish_oacc_routine (&data, NULL_TREE, false, true, false); - else - c_parser_declaration_or_fndef (parser, true, false, false, false, - true, NULL, vNULL, &data); + /* Emit a helpful diagnostic if there's another pragma following this + one. Also don't allow a static assertion declaration, as in the + following we'll just parse a *single* "declaration or function + definition", and the static assertion counts an one. */ + if (c_parser_next_token_is (parser, CPP_PRAGMA) + || c_parser_next_token_is_keyword (parser, RID_STATIC_ASSERT)) + { + error_at (data.loc, + "%<#pragma acc routine%> not immediately followed by" + " function declaration or definition"); + /* ..., and then just keep going. */ + return; + } + + /* We only have to consider the pragma_external case here. */ + if (c_parser_next_token_is (parser, CPP_KEYWORD) + && c_parser_peek_token (parser)->keyword == RID_EXTENSION) + { + int ext = disable_extension_diagnostics (); + do + c_parser_consume_token (parser); + while (c_parser_next_token_is (parser, CPP_KEYWORD) + && c_parser_peek_token (parser)->keyword == RID_EXTENSION); + c_parser_declaration_or_fndef (parser, true, true, true, false, true, + NULL, vNULL, &data); + restore_extension_diagnostics (ext); + } + else + c_parser_declaration_or_fndef (parser, true, true, true, false, true, + NULL, vNULL, &data); + } } /* Finalize an OpenACC routine pragma, applying it to FNDECL. @@ -14087,24 +14127,46 @@ c_parser_oacc_routine (c_parser *parser, enum pragma_context context) static void c_finish_oacc_routine (struct oacc_routine_data *data, tree fndecl, - bool named, bool first, bool is_defn) + bool is_defn) { - if (!fndecl || TREE_CODE (fndecl) != FUNCTION_DECL || !first) + /* Keep going if we're in error reporting mode. */ + if (data->error_seen + || fndecl == error_mark_node) + return; + + if (data->fndecl_seen) { - if (fndecl != error_mark_node) - error_at (data->loc, "%<#pragma acc routine%> %s", - named ? "does not refer to a function" - : "not followed by single function"); + error_at (data->loc, + "%<#pragma acc routine%> not immediately followed by" + " a single function declaration or definition"); + data->error_seen = true; + return; + } + if (fndecl == NULL_TREE || TREE_CODE (fndecl) != FUNCTION_DECL) + { + error_at (data->loc, + "%<#pragma acc routine%> not immediately followed by" + " function declaration or definition"); + data->error_seen = true; return; } if (get_oacc_fn_attrib (fndecl)) - error_at (data->loc, - "%<#pragma acc routine%> already applied to %D", fndecl); + { + error_at (data->loc, + "%<#pragma acc routine%> already applied to %qD", fndecl); + data->error_seen = true; + return; + } if (TREE_USED (fndecl) || (!is_defn && DECL_SAVED_TREE (fndecl))) - error_at (data->loc, "%<#pragma acc routine%> must be applied before %s", - TREE_USED (fndecl) ? "use" : "definition"); + { + error_at (data->loc, + "%<#pragma acc routine%> must be applied before %s", + TREE_USED (fndecl) ? "use" : "definition"); + data->error_seen = true; + return; + } /* Process the routine's dimension clauses. */ tree dims = build_oacc_routine_dims (data->clauses); @@ -14114,6 +14176,9 @@ c_finish_oacc_routine (struct oacc_routine_data *data, tree fndecl, DECL_ATTRIBUTES (fndecl) = tree_cons (get_identifier ("omp declare target"), NULL_TREE, DECL_ATTRIBUTES (fndecl)); + + /* Remember that we've used this "#pragma acc routine". */ + data->fndecl_seen = true; } /* OpenACC 2.0: diff --git a/gcc/cp/ChangeLog b/gcc/cp/ChangeLog index 45e647dd1f7..4c2b1157dfd 100644 --- a/gcc/cp/ChangeLog +++ b/gcc/cp/ChangeLog @@ -1,5 +1,13 @@ 2016-08-04 Thomas Schwinge + * parser.c (cp_ensure_no_oacc_routine): Improve diagnostics. + (cp_parser_late_parsing_cilk_simd_fn_info): Fix diagnostics. + (cp_parser_late_parsing_oacc_routine, cp_finalize_oacc_routine): + Simplify code, and improve diagnostics. + (cp_parser_oacc_routine): Likewise. Move pragma context + checking... + (cp_parser_pragma): ... here. + * parser.h (struct cp_omp_declare_simd_data): New. (struct cp_parser): Use it for oacc_routine member. * parser.c (cp_ensure_no_oacc_routine, cp_parser_oacc_routine) diff --git a/gcc/cp/parser.c b/gcc/cp/parser.c index e696d465192..fdb0ab0b035 100644 --- a/gcc/cp/parser.c +++ b/gcc/cp/parser.c @@ -1392,8 +1392,8 @@ cp_ensure_no_oacc_routine (cp_parser *parser) if (parser->oacc_routine && !parser->oacc_routine->error_seen) { error_at (parser->oacc_routine->loc, - "%<#pragma acc routine%> not followed by a function " - "declaration or definition"); + "%<#pragma acc routine%> not immediately followed by " + "function declaration or definition"); parser->oacc_routine = NULL; } } @@ -35736,6 +35736,8 @@ cp_parser_omp_declare_simd (cp_parser *parser, cp_token *pragma_tok, used while this scope is live. */ parser->omp_declare_simd = &data; } + + /* Store away all pragma tokens. */ while (cp_lexer_next_token_is_not (parser->lexer, CPP_PRAGMA_EOL) && cp_lexer_next_token_is_not (parser->lexer, CPP_EOF)) cp_lexer_consume_token (parser->lexer); @@ -35745,6 +35747,7 @@ cp_parser_omp_declare_simd (cp_parser *parser, cp_token *pragma_tok, struct cp_token_cache *cp = cp_token_cache_new (pragma_tok, cp_lexer_peek_token (parser->lexer)); parser->omp_declare_simd->tokens.safe_push (cp); + if (first_p) { while (cp_lexer_next_token_is (parser->lexer, CPP_PRAGMA)) @@ -35789,9 +35792,9 @@ cp_parser_late_parsing_cilk_simd_fn_info (cp_parser *parser, tree attrs) if (parser->omp_declare_simd != NULL || lookup_attribute ("simd", attrs)) { - error ("%<#pragma omp declare simd%> of % attribute cannot be " + error ("%<#pragma omp declare simd%> or % attribute cannot be " "used in the same function marked as a Cilk Plus SIMD-enabled " - " function"); + "function"); parser->cilk_simd_fn_info->tokens.release (); XDELETE (parser->cilk_simd_fn_info); parser->cilk_simd_fn_info = NULL; @@ -36563,64 +36566,39 @@ static void cp_parser_oacc_routine (cp_parser *parser, cp_token *pragma_tok, enum pragma_context context) { - bool first_p = parser->oacc_routine == NULL; - cp_oacc_routine_data data; - if (first_p) - { - data.error_seen = false; - data.fndecl_seen = false; - data.tokens = vNULL; - data.clauses = NULL_TREE; - data.loc = pragma_tok->location; - /* It is safe to take the address of a local variable; it will only be - used while this scope is live. */ - parser->oacc_routine = &data; - } + gcc_checking_assert (context == pragma_external); + /* The checking for "another pragma following this one" in the "no optional + '( name )'" case makes sure that we dont re-enter. */ + gcc_checking_assert (parser->oacc_routine == NULL); - if (context != pragma_external) - { - cp_parser_error (parser, "%<#pragma acc routine%> not at file scope"); - parser->oacc_routine->error_seen = true; - parser->oacc_routine = NULL; - return; - } + cp_oacc_routine_data data; + data.error_seen = false; + data.fndecl_seen = false; + data.tokens = vNULL; + data.clauses = NULL_TREE; + data.loc = pragma_tok->location; + /* It is safe to take the address of a local variable; it will only be + used while this scope is live. */ + parser->oacc_routine = &data; /* Look for optional '( name )'. */ if (cp_lexer_next_token_is (parser->lexer, CPP_OPEN_PAREN)) { - if (!first_p) - { - while (cp_lexer_next_token_is_not (parser->lexer, CPP_PRAGMA_EOL) - && cp_lexer_next_token_is_not (parser->lexer, CPP_EOF)) - cp_lexer_consume_token (parser->lexer); - if (cp_lexer_next_token_is_not (parser->lexer, CPP_PRAGMA_EOL)) - parser->oacc_routine->error_seen = true; - cp_parser_require_pragma_eol (parser, pragma_tok); - - error_at (parser->oacc_routine->loc, - "%<#pragma acc routine%> not followed by a " - "function declaration or definition"); - - parser->oacc_routine->error_seen = true; - return; - } - - cp_lexer_consume_token (parser->lexer); - cp_token *token = cp_lexer_peek_token (parser->lexer); + cp_lexer_consume_token (parser->lexer); /* '(' */ /* We parse the name as an id-expression. If it resolves to anything other than a non-overloaded function at namespace scope, it's an error. */ - tree id = cp_parser_id_expression (parser, - /*template_keyword_p=*/false, - /*check_dependency_p=*/false, - /*template_p=*/NULL, - /*declarator_p=*/false, - /*optional_p=*/false); - tree decl = cp_parser_lookup_name_simple (parser, id, token->location); - if (id != error_mark_node && decl == error_mark_node) - cp_parser_name_lookup_error (parser, id, decl, NLE_NULL, - token->location); + location_t name_loc = cp_lexer_peek_token (parser->lexer)->location; + tree name = cp_parser_id_expression (parser, + /*template_keyword_p=*/false, + /*check_dependency_p=*/false, + /*template_p=*/NULL, + /*declarator_p=*/false, + /*optional_p=*/false); + tree decl = cp_parser_lookup_name_simple (parser, name, name_loc); + if (name != error_mark_node && decl == error_mark_node) + cp_parser_name_lookup_error (parser, name, decl, NLE_NULL, name_loc); if (decl == error_mark_node || !cp_parser_require (parser, CPP_CLOSE_PAREN, RT_CLOSE_PAREN)) @@ -36630,8 +36608,6 @@ cp_parser_oacc_routine (cp_parser *parser, cp_token *pragma_tok, return; } - /* Build a chain of clauses. */ - parser->lexer->in_pragma = true; data.clauses = cp_parser_oacc_all_clauses (parser, OACC_ROUTINE_CLAUSE_MASK, "#pragma acc routine", @@ -36641,7 +36617,7 @@ cp_parser_oacc_routine (cp_parser *parser, cp_token *pragma_tok, && (TREE_CODE (decl) != FUNCTION_DECL || DECL_FUNCTION_TEMPLATE_P (decl))) { - error_at (data.loc, + error_at (name_loc, "%<#pragma acc routine%> names a set of overloads"); parser->oacc_routine = NULL; return; @@ -36651,60 +36627,58 @@ cp_parser_oacc_routine (cp_parser *parser, cp_token *pragma_tok, namespaces? */ if (!DECL_NAMESPACE_SCOPE_P (decl)) { - error_at (data.loc, - "%<#pragma acc routine%> does not refer to a " - "namespace scope function"); + error_at (name_loc, + "%qD does not refer to a namespace scope function", decl); parser->oacc_routine = NULL; return; } - if (!decl || TREE_CODE (decl) != FUNCTION_DECL) + if (TREE_CODE (decl) != FUNCTION_DECL) { - error_at (data.loc, - "%<#pragma acc routine%> does not refer to a function"); + error_at (name_loc, "%qD does not refer to a function", decl); parser->oacc_routine = NULL; return; } cp_finalize_oacc_routine (parser, decl, false); - data.tokens.release (); parser->oacc_routine = NULL; } - else + else /* No optional '( name )'. */ { + /* Store away all pragma tokens. */ while (cp_lexer_next_token_is_not (parser->lexer, CPP_PRAGMA_EOL) && cp_lexer_next_token_is_not (parser->lexer, CPP_EOF)) cp_lexer_consume_token (parser->lexer); if (cp_lexer_next_token_is_not (parser->lexer, CPP_PRAGMA_EOL)) parser->oacc_routine->error_seen = true; cp_parser_require_pragma_eol (parser, pragma_tok); - struct cp_token_cache *cp = cp_token_cache_new (pragma_tok, cp_lexer_peek_token (parser->lexer)); parser->oacc_routine->tokens.safe_push (cp); - while (cp_lexer_next_token_is (parser->lexer, CPP_PRAGMA)) - cp_parser_pragma (parser, context, NULL); - - if (first_p) + /* Emit a helpful diagnostic if there's another pragma following this + one. */ + if (cp_lexer_next_token_is (parser->lexer, CPP_PRAGMA)) { - cp_parser_declaration (parser); - - if (parser->oacc_routine - && !parser->oacc_routine->error_seen - && !parser->oacc_routine->fndecl_seen) - error_at (data.loc, - "%<#pragma acc routine%> not followed by a " - "function declaration or definition"); - + cp_ensure_no_oacc_routine (parser); data.tokens.release (); - parser->oacc_routine = NULL; + /* ..., and then just keep going. */ + return; } + + /* We only have to consider the pragma_external case here. */ + cp_parser_declaration (parser); + if (parser->oacc_routine + && !parser->oacc_routine->fndecl_seen) + cp_ensure_no_oacc_routine (parser); + else + parser->oacc_routine = NULL; + data.tokens.release (); } } /* Finalize #pragma acc routine clauses after direct declarator has - been parsed, and put that into "oacc function" attribute. */ + been parsed. */ static tree cp_parser_late_parsing_oacc_routine (cp_parser *parser, tree attrs) @@ -36712,17 +36686,17 @@ cp_parser_late_parsing_oacc_routine (cp_parser *parser, tree attrs) struct cp_token_cache *ce; cp_oacc_routine_data *data = parser->oacc_routine; - if ((!data->error_seen && data->fndecl_seen) - || data->tokens.length () != 1) + if (!data->error_seen && data->fndecl_seen) { error_at (data->loc, - "%<#pragma acc routine%> not followed by a " - "function declaration or definition"); + "%<#pragma acc routine%> not immediately followed by " + "a single function declaration or definition"); data->error_seen = true; } if (data->error_seen) return attrs; + gcc_checking_assert (data->tokens.length () == 1); ce = data->tokens[0]; cp_parser_push_lexer_for_tokens (parser, ce); @@ -36730,12 +36704,14 @@ cp_parser_late_parsing_oacc_routine (cp_parser *parser, tree attrs) gcc_assert (cp_lexer_peek_token (parser->lexer)->type == CPP_PRAGMA); cp_token *pragma_tok = cp_lexer_consume_token (parser->lexer); + gcc_checking_assert (parser->oacc_routine->clauses == NULL_TREE); parser->oacc_routine->clauses = cp_parser_oacc_all_clauses (parser, OACC_ROUTINE_CLAUSE_MASK, "#pragma acc routine", pragma_tok); cp_parser_pop_lexer (parser); + /* Later, cp_finalize_oacc_routine will process the clauses, and then set + fndecl_seen. */ - data->fndecl_seen = true; return attrs; } @@ -36747,34 +36723,29 @@ cp_finalize_oacc_routine (cp_parser *parser, tree fndecl, bool is_defn) { if (__builtin_expect (parser->oacc_routine != NULL, 0)) { - if (parser->oacc_routine->error_seen) + /* Keep going if we're in error reporting mode. */ + if (parser->oacc_routine->error_seen + || fndecl == error_mark_node) return; - - if (fndecl == error_mark_node) + + if (parser->oacc_routine->fndecl_seen) { + error_at (parser->oacc_routine->loc, + "%<#pragma acc routine%> not immediately followed by" + " a single function declaration or definition"); parser->oacc_routine = NULL; return; } - if (TREE_CODE (fndecl) != FUNCTION_DECL) { cp_ensure_no_oacc_routine (parser); return; } - if (!fndecl || TREE_CODE (fndecl) != FUNCTION_DECL) - { - error_at (parser->oacc_routine->loc, - "%<#pragma acc routine%> not followed by a function " - "declaration or definition"); - parser->oacc_routine = NULL; - return; - } - if (get_oacc_fn_attrib (fndecl)) { error_at (parser->oacc_routine->loc, - "%<#pragma acc routine%> already applied to %D", fndecl); + "%<#pragma acc routine%> already applied to %qD", fndecl); parser->oacc_routine = NULL; return; } @@ -36796,6 +36767,11 @@ cp_finalize_oacc_routine (cp_parser *parser, tree fndecl, bool is_defn) DECL_ATTRIBUTES (fndecl) = tree_cons (get_identifier ("omp declare target"), NULL_TREE, DECL_ATTRIBUTES (fndecl)); + + /* Don't unset parser->oacc_routine here: we may still need it to + diagnose wrong usage. But, remember that we've used this "#pragma acc + routine". */ + parser->oacc_routine->fndecl_seen = true; } } @@ -37395,6 +37371,12 @@ cp_parser_pragma (cp_parser *parser, enum pragma_context context, bool *if_p) return false; case PRAGMA_OACC_ROUTINE: + if (context != pragma_external) + { + error_at (pragma_tok->location, + "%<#pragma acc routine%> must be at file scope"); + break; + } cp_parser_oacc_routine (parser, pragma_tok, context); return false; diff --git a/gcc/testsuite/ChangeLog b/gcc/testsuite/ChangeLog index 0b2a7f211a8..0d6a8d2007b 100644 --- a/gcc/testsuite/ChangeLog +++ b/gcc/testsuite/ChangeLog @@ -1,3 +1,7 @@ +2016-08-04 Thomas Schwinge + + * c-c++-common/goacc/routine-5.c: Update. + 2016-08-04 Bernd Edlinger PR rtl-optimization/70903 diff --git a/gcc/testsuite/c-c++-common/goacc/routine-5.c b/gcc/testsuite/c-c++-common/goacc/routine-5.c index 1efd154f8e6..17fe67cd298 100644 --- a/gcc/testsuite/c-c++-common/goacc/routine-5.c +++ b/gcc/testsuite/c-c++-common/goacc/routine-5.c @@ -1,64 +1,211 @@ -/* { dg-do compile } */ +/* Miscellaneous OpenACC routine front end checking. */ -#pragma acc routine /* { dg-error "not followed by" } */ +/* Pragma context. */ + +struct PC +{ +#pragma acc routine /* { dg-error ".#pragma acc routine. must be at file scope" } */ +}; + +void PC1( /* { dg-bogus "variable or field .PC1. declared void" "TODO" { xfail c++ } } */ +#pragma acc routine + /* { dg-error ".#pragma acc routine. must be at file scope" "" { target c } 11 } + { dg-error ".#pragma. is not allowed here" "" { target c++ } 11 } */ +) /* { dg-bogus "expected declaration specifiers or .\\.\\.\\.. before .\\). token" "TODO" { xfail c } } */ +{ +} + +void PC2() +{ + if (0) +#pragma acc routine /* { dg-error ".#pragma acc routine. must be at file scope" } */ + ; +} + +void PC3() +{ +#pragma acc routine /* { dg-error ".#pragma acc routine. must be at file scope" } */ +} + + +/* "( name )" syntax. */ + +#pragma acc routine ( /* { dg-error "expected (function name|unqualified-id) before end of line" } */ +#pragma acc routine () /* { dg-error "expected (function name|unqualified-id) before .\\). token" } */ +#pragma acc routine (+) /* { dg-error "expected (function name|unqualified-id) before .\\+. token" } */ +#pragma acc routine (?) /* { dg-error "expected (function name|unqualified-id) before .\\?. token" } */ +#pragma acc routine (:) /* { dg-error "expected (function name|unqualified-id) before .:. token" } */ +#pragma acc routine (4) /* { dg-error "expected (function name|unqualified-id) before numeric constant" } */ +#pragma acc routine ('4') /* { dg-error "expected (function name|unqualified-id) before .4." } */ +#pragma acc routine ("4") /* { dg-error "expected (function name|unqualified-id) before string constant" } */ +extern void R1(void); +extern void R2(void); +#pragma acc routine (R1, R2, R3) worker /* { dg-error "expected .\\). before .,. token" } */ +#pragma acc routine (R1 R2 R3) worker /* { dg-error "expected .\\). before .R2." } */ +#pragma acc routine (R1) worker +#pragma acc routine (R2) worker + + +/* "#pragma acc routine" not immediately followed by (a single) function + declaration or definition. */ + +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ int a; -#pragma acc routine /* { dg-error "not followed by" } */ +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by a single function declaration or definition" } */ void fn1 (void), fn1b (void); -#pragma acc routine /* { dg-error "not followed by" } */ +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ int b, fn2 (void); -#pragma acc routine /* { dg-error "not followed by" } */ +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ +int b_, fn2_ (void), B_; + +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by a single function declaration or definition" } */ int fn3 (void), b2; -#pragma acc routine /* { dg-error "not followed by" } */ +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ typedef struct c c; -#pragma acc routine /* { dg-error "not followed by" } */ +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ struct d {} d; -#pragma acc routine /* { dg-error "not followed by" } */ +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by a single function declaration or definition" } */ +void fn1_2 (void), fn1b_2 (void); + +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ +int b_2, fn2_2 (void); + +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ +int b_2_, fn2_2_ (void), B_2_; + +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by a single function declaration or definition" } */ +int fn3_2 (void), b2_2; + +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ +typedef struct c_2 c_2; + +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ +struct d_2 {} d_2; + +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ #pragma acc routine int fn4 (void); int fn5a (void); - -#pragma acc routine /* { dg-error "not followed by" } */ +int fn5b (void); +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ #pragma acc routine (fn5a) +#pragma acc routine (fn5b) int fn5 (void); +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ +#pragma acc routine (fn6a) /* { dg-error ".fn6a. has not been declared" } */ +#pragma acc routine (fn6b) /* { dg-error ".fn6b. has not been declared" } */ +int fn6 (void); + #ifdef __cplusplus -#pragma acc routine /* { dg-error "not followed by" "" { target c++ } } */ +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" "" { target c++ } } */ namespace f {} namespace g {} -#pragma acc routine /* { dg-error "not followed by" "" { target c++ } } */ +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" "" { target c++ } } */ using namespace g; -#pragma acc routine (g) /* { dg-error "does not refer to a function" "" { target c++ } } */ +#pragma acc routine (g) /* { dg-error ".g. does not refer to a function" "" { target c++ } } */ #endif /* __cplusplus */ -#pragma acc routine (a) /* { dg-error "does not refer to a function" } */ +#pragma acc routine (a) /* { dg-error ".a. does not refer to a function" } */ -#pragma acc routine (c) /* { dg-error "does not refer to a function" } */ +#pragma acc routine (c) /* { dg-error ".c. does not refer to a function" } */ -#pragma acc routine () vector /* { dg-error "expected (function name|unqualified-id) before .\\). token" } */ +/* Static assert. */ -#pragma acc routine (+) /* { dg-error "expected (function name|unqualified-id) before .\\+. token" } */ +#pragma acc routine /* { dg-bogus ".#pragma acc routine. not immediately followed by function declaration or definition" "TODO" { xfail *-*-* } } */ +#ifndef __cplusplus /* C */ +_Static_assert(0, ""); /* { dg-error "static assertion failed" "" { target c } } */ +#elif __cplusplus < 201103L /* C++98 */ +/* C++98 doesn't support static_assert, so fake an error in combination, and as + expected with the "#pragma acc routine" above. */ +int dummy_instead_of_static_assert; +#else /* C++ */ +static_assert(0, ""); /* { dg-error "static assertion failed" "" { target c++11 } } */ +#endif +void f_static_assert(); +/* Check that we already recognized "f_static_assert" as an OpenACC routine. */ +#pragma acc routine (f_static_assert) /* { dg-error ".#pragma acc routine. already applied to .\[void \]*f_static_assert" "TODO" { xfail *-*-* } } */ -extern void R1(void); -extern void R2(void); -#pragma acc routine (R1, R2, R3) worker /* { dg-error "expected .\\). before .,. token" } */ -#pragma acc routine (R1 R2 R3) worker /* { dg-error "expected .\\). before .R2." } */ -#pragma acc routine (R1) worker -#pragma acc routine (R2) worker +/* __extension__ usage. */ + +#pragma acc routine +__extension__ extern void ex1(); +#pragma acc routine (ex1) /* { dg-error ".#pragma acc routine. already applied to .\[void \]*ex1" } */ + +#pragma acc routine +__extension__ __extension__ __extension__ __extension__ __extension__ void ex2() +{ +} +#pragma acc routine (ex2) /* { dg-error ".#pragma acc routine. already applied to .\[void \]*ex2" } */ + +#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */ +__extension__ int ex3; +#pragma acc routine (ex3) /* { dg-error ".ex3. does not refer to a function" } */ + + +/* "#pragma acc routine" already applied. */ +extern void fungsi_1(); +#pragma acc routine(fungsi_1) gang +#pragma acc routine(fungsi_1) gang /* { dg-error ".#pragma acc routine. already applied to .\[void \]*fungsi_1" } */ +#pragma acc routine(fungsi_1) worker /* { dg-error ".#pragma acc routine. already applied to .\[void \]*fungsi_1" } */ +#pragma acc routine(fungsi_1) vector /* { dg-error ".#pragma acc routine. already applied to .\[void \]*fungsi_1" } */ + +#pragma acc routine seq +extern void fungsi_2(); +#pragma acc routine(fungsi_2) seq /* { dg-error ".#pragma acc routine. already applied to .\[void \]*fungsi_2." } */ +#pragma acc routine(fungsi_2) worker /* { dg-error ".#pragma acc routine. already applied to .\[void \]*fungsi_2." } */ +#pragma acc routine(fungsi_2) /* { dg-error ".#pragma acc routine. already applied to .\[void \]*fungsi_2." } */ + +#pragma acc routine vector +extern void fungsi_3(); +#pragma acc routine vector /* { dg-error ".#pragma acc routine. already applied to .\[void \]*fungsi_3." } */ +void fungsi_3() +{ +} + +extern void fungsi_4(); +#pragma acc routine (fungsi_4) worker +#pragma acc routine gang /* { dg-error ".#pragma acc routine. already applied to .\[void \]*fungsi_4." } */ +void fungsi_4() +{ +} + +#pragma acc routine gang +void fungsi_5() +{ +} +#pragma acc routine (fungsi_5) worker /* { dg-error ".#pragma acc routine. already applied to .\[void \]*fungsi_5." } */ + +#pragma acc routine seq +void fungsi_6() +{ +} +#pragma acc routine seq /* { dg-error ".#pragma acc routine. already applied to .\[void \]*fungsi_6." } */ +extern void fungsi_6(); + + +/* "#pragma acc routine" must be applied before. */ void Bar (); @@ -67,13 +214,15 @@ void Foo () Bar (); } -#pragma acc routine (Bar) // { dg-error "must be applied before use" } +#pragma acc routine (Bar) // { dg-error ".#pragma acc routine. must be applied before use" } -#pragma acc routine (Foo) gang // { dg-error "must be applied before definition" } +#pragma acc routine (Foo) gang // { dg-error ".#pragma acc routine. must be applied before definition" } #pragma acc routine (Baz) // { dg-error "not been declared" } +/* OpenACC declare. */ + int vb1; /* { dg-error "directive for use" } */ extern int vb2; /* { dg-error "directive for use" } */ static int vb3; /* { dg-error "directive for use" } */ -- 2.30.2