Hi!
Ping.
On Wed, 13 Jul 2016 16:10:31 +0200, I wrote:
> On Wed, 13 Jul 2016 11:25:46 +0200, I wrote:
> > Working on something else regarding the C/C++ OpenACC routine directive,
> > I couldn't but untangle [...]
>
> > (Another C/C++ OpenACC routine
> > cleanup patch is emerging, depending on this one.)
>
> Here it is; likewise, OK for trunk? (Further cleanup especially of C++
> OpenACC routine handling seems to be possible, but I want to synchronize
> my work at this point.)
>
> commit 0bd30acaf4dd634499b1c695ddee555e7675aa18
> Author: Thomas Schwinge <tho...@codesourcery.com>
> Date: Thu Jun 23 13:28:09 2016 +0200
>
> 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.
> ---
> gcc/c/c-parser.c | 161 +++++++++++++++-------
> gcc/cp/parser.c | 182 +++++++++++-------------
> gcc/testsuite/c-c++-common/goacc/routine-5.c | 199
> +++++++++++++++++++++++----
> 3 files changed, 369 insertions(+), 173 deletions(-)
>
> diff --git gcc/c/c-parser.c gcc/c/c-parser.c
> index 7f84ce9..809118a 100644
> --- gcc/c/c-parser.c
> +++ gcc/c/c-parser.c
> @@ -1273,6 +1273,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;
> };
> @@ -1565,8 +1567,7 @@ c_parser_external_declaration (c_parser *parser)
> }
>
> static void c_finish_omp_declare_simd (c_parser *, tree, tree, vec<c_token>);
> -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
> @@ -1751,8 +1752,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;
> }
>
> @@ -1850,7 +1850,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;
> @@ -1870,8 +1870,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;
> }
> @@ -1987,8 +1986,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);
> @@ -2033,8 +2031,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);
> @@ -2146,8 +2143,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);
> @@ -10117,6 +10113,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;
>
> @@ -14029,29 +14032,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 )'. */
> + /* Look for optional '( name )'. */
> if (c_parser_peek_token (parser)->type == 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
> @@ -14063,22 +14069,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");
> + /* 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;
> + }
>
> - 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);
> + /* 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.
> @@ -14086,24 +14126,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);
> @@ -14113,6 +14175,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 gcc/cp/parser.c gcc/cp/parser.c
> index 28417d8..43bed89 100644
> --- gcc/cp/parser.c
> +++ gcc/cp/parser.c
> @@ -1383,8 +1383,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;
> }
> }
> @@ -35660,6 +35660,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);
> @@ -35669,6 +35671,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))
> @@ -35713,9 +35716,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 %<simd%> attribute cannot be "
> + error ("%<#pragma omp declare simd%> or %<simd%> 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;
> @@ -36487,64 +36490,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;
> + 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);
> +
> 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;
> - }
> + 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;
>
> - 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;
> - }
> -
> - /* Look for optional '( name )'. */
> + /* Do we have an 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))
> @@ -36554,8 +36532,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",
> @@ -36565,7 +36541,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;
> @@ -36575,60 +36551,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)
> @@ -36636,17 +36610,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);
> @@ -36654,12 +36628,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;
> }
>
> @@ -36671,34 +36647,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;
> }
> @@ -36720,6 +36691,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;
> }
> }
>
> @@ -37319,6 +37295,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 gcc/testsuite/c-c++-common/goacc/routine-5.c
> gcc/testsuite/c-c++-common/goacc/routine-5.c
> index 1efd154..17fe67c 100644
> --- gcc/testsuite/c-c++-common/goacc/routine-5.c
> +++ 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" } */
Grüße
Thomas
