From patchwork Mon Aug 17 17:47:24 2015 Content-Type: text/plain; charset="utf-8" MIME-Version: 1.0 Content-Transfer-Encoding: 7bit X-Patchwork-Submitter: Nathan Sidwell X-Patchwork-Id: 508044 Return-Path: X-Original-To: incoming@patchwork.ozlabs.org Delivered-To: patchwork-incoming@bilbo.ozlabs.org Received: from sourceware.org (server1.sourceware.org [209.132.180.131]) (using TLSv1.2 with cipher ECDHE-RSA-AES256-GCM-SHA384 (256/256 bits)) (No client certificate requested) by ozlabs.org (Postfix) with ESMTPS id 5581314012C for ; Tue, 18 Aug 2015 03:47:40 +1000 (AEST) Authentication-Results: ozlabs.org; dkim=pass (1024-bit key; unprotected) header.d=gcc.gnu.org header.i=@gcc.gnu.org header.b=Gk+xSIB+; dkim-atps=neutral DomainKey-Signature: a=rsa-sha1; c=nofws; d=gcc.gnu.org; h=list-id :list-unsubscribe:list-archive:list-post:list-help:sender:to :from:subject:message-id:date:mime-version:content-type; q=dns; s=default; b=W2dP2nzPfsobBvnAaZVQ6PM8jgmUqVHzYkFj3S+cdgLZTPfrV+ NDnVm2rc9P0FX2rJ3xhPkbX9mbHAV/hQsx2w5dE8aivDjJTVUTHUZ982ViWhyic2 nn1TmqLM0ffp5Ad5ReYjH5cWMFsAwPvDnUBzlmD9zWEm9/Q2CUGvvOzr4= DKIM-Signature: v=1; a=rsa-sha1; c=relaxed; d=gcc.gnu.org; h=list-id :list-unsubscribe:list-archive:list-post:list-help:sender:to :from:subject:message-id:date:mime-version:content-type; s= default; bh=86k78c+G/Icb5otkEkwsrw7BkfU=; b=Gk+xSIB+Sleh/nno/X+L ZTDq2bcju1E22ZZK9mqOgbDXhMZ8m5TlMTPxt7SC/83cGPmDL3w690BCu1Eren49 A6wGxljkt0z/RBkvO9qRjvRH1TKDVKaLuP0VDHb7I1aEJ7kaQkusCvIpmZpHTJpe nh6qnknDucZ+lATDju3g+iM= Received: (qmail 89416 invoked by alias); 17 Aug 2015 17:47:33 -0000 Mailing-List: contact gcc-patches-help@gcc.gnu.org; run by ezmlm Precedence: bulk List-Id: List-Unsubscribe: List-Archive: List-Post: List-Help: Sender: gcc-patches-owner@gcc.gnu.org Delivered-To: mailing list gcc-patches@gcc.gnu.org Received: (qmail 89404 invoked by uid 89); 17 Aug 2015 17:47:32 -0000 Authentication-Results: sourceware.org; auth=none X-Virus-Found: No X-Spam-SWARE-Status: No, score=0.9 required=5.0 tests=BAYES_50, FREEMAIL_FROM, KAM_ASCII_DIVIDERS, RCVD_IN_DNSWL_LOW, SPF_PASS autolearn=no version=3.3.2 X-HELO: mail-qg0-f46.google.com Received: from mail-qg0-f46.google.com (HELO mail-qg0-f46.google.com) (209.85.192.46) by sourceware.org (qpsmtpd/0.93/v0.84-503-g423c35a) with (AES128-GCM-SHA256 encrypted) ESMTPS; Mon, 17 Aug 2015 17:47:29 +0000 Received: by qgj62 with SMTP id 62so99275373qgj.2 for ; Mon, 17 Aug 2015 10:47:27 -0700 (PDT) X-Received: by 10.140.150.213 with SMTP id 204mr4689849qhw.104.1439833646973; Mon, 17 Aug 2015 10:47:26 -0700 (PDT) Received: from ?IPv6:2601:181:c000:c497:a2a8:cdff:fe3e:b48? ([2601:181:c000:c497:a2a8:cdff:fe3e:b48]) by smtp.googlemail.com with ESMTPSA id o66sm4063043qko.33.2015.08.17.10.47.25 (version=TLSv1.2 cipher=ECDHE-RSA-AES128-GCM-SHA256 bits=128/128); Mon, 17 Aug 2015 10:47:25 -0700 (PDT) To: GCC Patches From: Nathan Sidwell Subject: [gomp4] CXX parsing of routine Message-ID: <55D21E2C.4070907@acm.org> Date: Mon, 17 Aug 2015 13:47:24 -0400 User-Agent: Mozilla/5.0 (X11; Linux x86_64; rv:38.0) Gecko/20100101 Thunderbird/38.1.0 MIME-Version: 1.0 I've applied this patch to fix the C++ parsing of the routine directive. As with the C parser, this was constructing a list of names to apply the directive to, rather than simply resolving the name in the scope of the directive. With C++ this is even more interesting than with C, because names are not simple identifiers, but may involve scope operators. I also added checking to the C parser for use/defn before application, as well as duplicate application. testcases coming up next ... nathan 2015-08-17 Nathan Sidwell c/ * c-parser.c (c_finish_oacc_routine): Add is_defn arg and adjust all callers. Detect duplicate, post-use or post-defn application. cp/ * parser.h (struct cp_parser): Change oacc_routine field tree. Remove named_oacc_routines field. * parser.c (cp_ensure_no_omp_declare_simd): Remove oacc routine checking, move to ... (cp_ensure_no_oacc_routine): ... here. New function, adjust all callers. (cp_parser_late_parsing_oacc_routine): Delete. (cp_parser_new): Adjust. (cp_parser_linkage_specification): Call cp_ensure_no_oacc_routine. (cp_finalize_oacc_routine): Reimplement. Adjust all callers. (cp_parser_late_return_type_opt): Remove oacc routine handling. (cp_parser_omp_declare_simd): Remove oacc routine handling. Adjust callers. (cp_parser_finish_oacc_routine): New. (cp_parser_oacc_routine): Reimplement. testsuite/ * c-c++-common/goacc/routine-2.c: Insert declaration. libgomp/ * testsuite/libgomp.c-c++-common/routine-2.c: Insert declaration. Index: gcc/c/c-parser.c =================================================================== --- gcc/c/c-parser.c (revision 226912) +++ gcc/c/c-parser.c (working copy) @@ -1762,7 +1762,7 @@ finish_oacc_declare (tree fnbody, tree d static void c_finish_omp_declare_simd (c_parser *, tree, tree, vec); -static void c_finish_oacc_routine (c_parser *, tree, tree); +static void c_finish_oacc_routine (c_parser *, tree, 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 @@ -2020,7 +2020,8 @@ c_parser_declaration_or_fndef (c_parser c_finish_omp_declare_simd (parser, NULL_TREE, NULL_TREE, omp_declare_simd_clauses); if (oacc_routine_clauses) - c_finish_oacc_routine (parser, NULL_TREE, oacc_routine_clauses); + c_finish_oacc_routine (parser, NULL_TREE, + oacc_routine_clauses, false); c_parser_skip_to_end_of_block_or_statement (parser); return; } @@ -2117,9 +2118,6 @@ c_parser_declaration_or_fndef (c_parser || !vec_safe_is_empty (parser->cilk_simd_fn_tokens)) c_finish_omp_declare_simd (parser, d, NULL_TREE, omp_declare_simd_clauses); - - if (oacc_routine_clauses) - c_finish_oacc_routine (parser, d, oacc_routine_clauses); } else { @@ -2133,14 +2131,14 @@ c_parser_declaration_or_fndef (c_parser || !vec_safe_is_empty (parser->cilk_simd_fn_tokens)) c_finish_omp_declare_simd (parser, d, NULL_TREE, omp_declare_simd_clauses); - if (oacc_routine_clauses) - c_finish_oacc_routine (parser, d, oacc_routine_clauses); - + start_init (d, asm_name, global_bindings_p ()); init_loc = c_parser_peek_token (parser)->location; init = c_parser_initializer (parser); finish_init (); } + if (oacc_routine_clauses) + c_finish_oacc_routine (parser, d, oacc_routine_clauses, false); if (d != error_mark_node) { maybe_warn_string_init (init_loc, TREE_TYPE (d), init); @@ -2186,8 +2184,8 @@ c_parser_declaration_or_fndef (c_parser temp_pop_parm_decls (); } if (oacc_routine_clauses) - c_finish_oacc_routine (parser, d, oacc_routine_clauses); - + c_finish_oacc_routine (parser, d, oacc_routine_clauses, false); + if (d) finish_decl (d, UNKNOWN_LOCATION, NULL_TREE, NULL_TREE, asm_name); @@ -2298,10 +2296,10 @@ c_parser_declaration_or_fndef (c_parser || !vec_safe_is_empty (parser->cilk_simd_fn_tokens)) c_finish_omp_declare_simd (parser, current_function_decl, NULL_TREE, omp_declare_simd_clauses); - if (oacc_routine_clauses) c_finish_oacc_routine (parser, current_function_decl, - oacc_routine_clauses); + oacc_routine_clauses, true); + DECL_STRUCT_FUNCTION (current_function_decl)->function_start_locus = c_parser_peek_token (parser)->location; @@ -13279,6 +13277,10 @@ c_parser_oacc_parallel (location_t loc, | (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_SEQ) \ | (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_BIND)) +/* Parse an OpenACC routine directive. For named directives, we apply + immediately to the named function. For unnamed ones we then parse + a declaration or definition, which must be for a function. */ + static void c_parser_oacc_routine (c_parser *parser, enum pragma_context context) { @@ -13325,24 +13327,36 @@ c_parser_oacc_routine (c_parser *parser, clauses = tree_cons (c_head, clauses, NULL_TREE); if (decl) - c_finish_oacc_routine (parser, decl, clauses); + c_finish_oacc_routine (parser, decl, clauses, false); else c_parser_declaration_or_fndef (parser, true, false, false, false, true, NULL, vNULL, clauses); } +/* Finalize an OpenACC routine pragma, applying it to FNDECL. CLAUSES + are the parsed clauses. IS_DEFN is true if we're applying it to + the definition (so expect FNDEF to look somewhat defined. */ + static void c_finish_oacc_routine (c_parser *ARG_UNUSED (parser), - tree fndecl, tree clauses) + tree fndecl, tree clauses, bool is_defn) { + location_t loc = OMP_CLAUSE_LOCATION (TREE_PURPOSE (clauses)); + if (!fndecl || TREE_CODE (fndecl) != FUNCTION_DECL) { if (fndecl != error_mark_node) - error_at (OMP_CLAUSE_LOCATION (TREE_PURPOSE (clauses)), - "%<#pragma acc routine%> does not refer to a function"); + error_at (loc, "%<#pragma acc routine%> does not refer to a function"); return; } + if (get_oacc_fn_attrib (fndecl)) + error_at (loc, "%<#pragma acc routine%> already applied to %D", fndecl); + + if (TREE_USED (fndecl) || (!is_defn && DECL_SAVED_TREE (fndecl))) + error_at (loc, "%<#pragma acc routine%> must be applied before %s", + TREE_USED (fndecl) ? "use" : "definition"); + /* Process for function attrib */ tree dims = build_oacc_routine_dims (TREE_VALUE (clauses)); replace_oacc_fn_attrib (fndecl, dims); Index: gcc/cp/parser.c =================================================================== --- gcc/cp/parser.c (revision 226912) +++ gcc/cp/parser.c (working copy) @@ -251,6 +251,7 @@ static bool cp_parser_omp_declare_reduct (tree, cp_parser *); static tree cp_parser_cilk_simd_vectorlength (cp_parser *, tree, bool); +static void cp_finalize_oacc_routine (cp_parser *, tree, bool); /* Manifest constants. */ #define CP_LEXER_BUFFER_SIZE ((256 * 1024) / sizeof (cp_token)) @@ -1292,8 +1293,7 @@ cp_token_cache_new (cp_token *first, cp_ } /* Diagnose if #pragma omp declare simd isn't followed immediately - by function declaration or definition. Likewise for - #pragma acc routine. */ + by function declaration or definition. */ static inline void cp_ensure_no_omp_declare_simd (cp_parser *parser) @@ -1304,13 +1304,6 @@ cp_ensure_no_omp_declare_simd (cp_parser "function declaration or definition"); parser->omp_declare_simd = NULL; } - - if (parser->oacc_routine && !parser->oacc_routine->error_seen) - { - error ("%<#pragma acc routine%> not immediately followed by " - "function declaration or definition"); - parser->oacc_routine = NULL; - } } /* Finalize #pragma omp declare simd clauses after FNDECL has been parsed, @@ -1334,56 +1327,13 @@ cp_finalize_omp_declare_simd (cp_parser } } -/* Finalize #pragma acc routine clauses after FNDECL has been parsed, - and put that into "acc routine" attribute. */ +/* Diagnose if #pragma omp routine isn't followed immediately + by function declaration or definition. */ static inline void -cp_finalize_oacc_routine (cp_parser *parser, tree fndecl) +cp_ensure_no_oacc_routine (cp_parser *parser) { - if (__builtin_expect (parser->omp_declare_simd != NULL, 0)) - { - if (fndecl == error_mark_node) - { - parser->omp_declare_simd = NULL; - return; - } - if (TREE_CODE (fndecl) != FUNCTION_DECL) - { - cp_ensure_no_omp_declare_simd (parser); - return; - } - } - else // Is this fndecl associated with a named routine? - { - if (fndecl == NULL_TREE || fndecl == error_mark_node - || TREE_CODE (fndecl) != FUNCTION_DECL) - return; - - bool found = false; - int i; - tree t, clauses = NULL_TREE; - - for (i = 0; vec_safe_iterate (parser->named_oacc_routines, i, &t); i++) - { - if (!strcmp (IDENTIFIER_POINTER (DECL_NAME (fndecl)), - IDENTIFIER_POINTER (t))) - { - found = true; - clauses = TREE_CHAIN (t); - break; - } - } - - if (!found) - return; - - if (clauses != NULL_TREE) - clauses = tree_cons (NULL_TREE, clauses, NULL_TREE); - clauses = build_tree_list (get_identifier ("omp declare target"), - clauses); - TREE_CHAIN (clauses) = DECL_ATTRIBUTES (fndecl); - DECL_ATTRIBUTES (fndecl) = clauses; - } + cp_finalize_oacc_routine (parser, NULL_TREE, false); } /* Decl-specifiers. */ @@ -2244,9 +2194,6 @@ static tree cp_parser_late_parsing_omp_d static tree cp_parser_late_parsing_cilk_simd_fn_info (cp_parser *, tree); -static tree cp_parser_late_parsing_oacc_routine - (cp_parser *, tree); - static tree synthesize_implicit_template_parm (cp_parser *, tree); static tree finish_fully_implicit_template @@ -3672,9 +3619,8 @@ cp_parser_new (void) parser->implicit_template_parms = 0; parser->implicit_template_scope = 0; - /* The list of OpenACC routines pragmas is unitialized. */ + /* Active OpenACC routine clauses. */ parser->oacc_routine = NULL; - parser->named_oacc_routines = NULL; /* Allow constrained-type-specifiers. */ parser->prevent_constrained_type_specifiers = 0; @@ -12299,6 +12245,7 @@ cp_parser_linkage_specification (cp_pars if (cp_lexer_next_token_is (parser->lexer, CPP_OPEN_BRACE)) { cp_ensure_no_omp_declare_simd (parser); + cp_ensure_no_oacc_routine (parser); /* Consume the `{' token. */ cp_lexer_consume_token (parser->lexer); @@ -16807,6 +16754,7 @@ cp_parser_namespace_definition (cp_parse bool is_inline; cp_ensure_no_omp_declare_simd (parser); + cp_ensure_no_oacc_routine (parser); if (cp_lexer_next_token_is_keyword (parser->lexer, RID_INLINE)) { maybe_warn_cpp0x (CPP0X_INLINE_NAMESPACES); @@ -17787,7 +17735,7 @@ cp_parser_init_declarator (cp_parser* pa range_for_decl_p? SD_INITIALIZED : is_initialized, attributes, prefix_attributes, &pushed_scope); cp_finalize_omp_declare_simd (parser, decl); - cp_finalize_oacc_routine (parser, decl); + cp_finalize_oacc_routine (parser, decl, false); /* Adjust location of decl if declarator->id_loc is more appropriate: set, and decl wasn't merged with another decl, in which case its location would be different from input_location, and more accurate. */ @@ -17901,7 +17849,7 @@ cp_parser_init_declarator (cp_parser* pa if (decl && TREE_CODE (decl) == FUNCTION_DECL) cp_parser_save_default_args (parser, decl); cp_finalize_omp_declare_simd (parser, decl); - cp_finalize_oacc_routine (parser, decl); + cp_finalize_oacc_routine (parser, decl, false); } /* Finish processing the declaration. But, skip member @@ -18970,15 +18918,12 @@ cp_parser_late_return_type_opt (cp_parse bool cilk_simd_fn_vector_p = (parser->cilk_simd_fn_info && declarator && declarator->kind == cdk_id); - bool oacc_routine_p = (parser->oacc_routine - && declarator && declarator->kind == cdk_id); - /* Peek at the next token. */ token = cp_lexer_peek_token (parser->lexer); /* A late-specified return type is indicated by an initial '->'. */ if (token->type != CPP_DEREF && token->keyword != RID_REQUIRES - && !(declare_simd_p || cilk_simd_fn_vector_p || oacc_routine_p)) + && !(declare_simd_p || cilk_simd_fn_vector_p)) return NULL_TREE; tree save_ccp = current_class_ptr; @@ -19009,11 +18954,6 @@ cp_parser_late_return_type_opt (cp_parse declarator->std_attributes = cp_parser_late_parsing_omp_declare_simd (parser, declarator->std_attributes); - if (oacc_routine_p) - declarator->std_attributes - = cp_parser_late_parsing_oacc_routine (parser, - declarator->std_attributes); - if (quals >= 0) { current_class_ptr = save_ccp; @@ -20469,6 +20409,7 @@ cp_parser_class_specifier_1 (cp_parser* } cp_ensure_no_omp_declare_simd (parser); + cp_ensure_no_oacc_routine (parser); /* Issue an error message if type-definitions are forbidden here. */ cp_parser_check_type_definition (parser); @@ -21782,7 +21723,7 @@ cp_parser_member_declaration (cp_parser* } cp_finalize_omp_declare_simd (parser, decl); - cp_finalize_oacc_routine (parser, decl); + cp_finalize_oacc_routine (parser, decl, false); /* Reset PREFIX_ATTRIBUTES. */ while (attributes && TREE_CHAIN (attributes) != first_attribute) @@ -24390,8 +24331,7 @@ cp_parser_function_definition_from_speci cp_finalize_omp_declare_simd (parser, current_function_decl); parser->omp_declare_simd = NULL; - cp_finalize_oacc_routine (parser, current_function_decl); - parser->oacc_routine = NULL; + cp_finalize_oacc_routine (parser, current_function_decl, true); } if (!success_p) @@ -25074,7 +25014,7 @@ cp_parser_save_member_function_body (cp_ /* Create the FUNCTION_DECL. */ fn = grokmethod (decl_specifiers, declarator, attributes); cp_finalize_omp_declare_simd (parser, fn); - cp_finalize_oacc_routine (parser, fn); + cp_finalize_oacc_routine (parser, fn, true); /* If something went badly wrong, bail out now. */ if (fn == error_mark_node) { @@ -33693,41 +33633,29 @@ cp_parser_oacc_wait (cp_parser *parser, static void cp_parser_omp_declare_simd (cp_parser *parser, cp_token *pragma_tok, - enum pragma_context context, bool is_omp) + enum pragma_context context) { - bool first_p = is_omp ? parser->omp_declare_simd == NULL - : parser->oacc_routine == NULL; + bool first_p = parser->omp_declare_simd == NULL; cp_omp_declare_simd_data data; if (first_p) { data.error_seen = false; data.fndecl_seen = false; data.tokens = vNULL; - if (is_omp) - parser->omp_declare_simd = &data; - else - parser->oacc_routine = &data; + parser->omp_declare_simd = &data; } 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)) - { - if (is_omp) - parser->omp_declare_simd->error_seen = true; - else - parser->oacc_routine->error_seen = true; - } + parser->omp_declare_simd->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)); - if (is_omp) - parser->omp_declare_simd->tokens.safe_push (cp); - else - parser->oacc_routine->tokens.safe_push (cp); + parser->omp_declare_simd->tokens.safe_push (cp); if (first_p) { @@ -33748,23 +33676,14 @@ cp_parser_omp_declare_simd (cp_parser *p cp_parser_declaration_statement (parser); break; } - if (is_omp && parser->omp_declare_simd + if (parser->omp_declare_simd && !parser->omp_declare_simd->error_seen && !parser->omp_declare_simd->fndecl_seen) error_at (pragma_tok->location, "%<#pragma omp declare simd%> not immediately followed by " "function declaration or definition"); - else if (!is_omp && parser->oacc_routine - && !parser->oacc_routine->error_seen - && !parser->oacc_routine->fndecl_seen) - error_at (pragma_tok->location, - "%<#pragma acc routine%> not immediately followed by " - "function declaration or definition"); data.tokens.release (); - if (is_omp) - parser->omp_declare_simd = NULL; - else - parser->oacc_routine = NULL; + parser->omp_declare_simd = NULL; } } @@ -34339,8 +34258,7 @@ cp_parser_omp_declare (cp_parser *parser if (strcmp (p, "simd") == 0) { cp_lexer_consume_token (parser->lexer); - cp_parser_omp_declare_simd (parser, pragma_tok, - context, true); + cp_parser_omp_declare_simd (parser, pragma_tok, context); return; } cp_ensure_no_omp_declare_simd (parser); @@ -34391,42 +34309,103 @@ cp_parser_omp_declare (cp_parser *parser | (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_SEQ) \ | (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_BIND)) +/* Finalize #pragma acc routine clauses after direct declarator has + been parsed, and put that into "omp declare target" attribute. */ + +static void +cp_parser_finish_oacc_routine (cp_parser *ARG_UNUSED (parser), tree fndecl, + tree clauses, bool named, bool is_defn) +{ + location_t loc = OMP_CLAUSE_LOCATION (TREE_PURPOSE (clauses)); + + if (!fndecl) + { + error ("%<#pragma oacc routine%> not immediately followed by " + "function declaration or definition"); + return; + } + + if (named && is_overloaded_fn (fndecl) + && (TREE_CODE (fndecl) != FUNCTION_DECL + || DECL_FUNCTION_TEMPLATE_P (fndecl))) + { + error_at (loc, "%D names a set of overloads", OVL_CURRENT (fndecl)); + return; + } + + if (TREE_CODE (fndecl) != FUNCTION_DECL) + { + error_at (loc, "%D is not a function", fndecl); + return; + } + + /* Perhaps we should use the same rule as declarations in different + namespaces? */ + if (named && !DECL_NAMESPACE_SCOPE_P (fndecl)) + { + error_at (loc, "%D is not at namespace scope", fndecl); + return; + } + + if (get_oacc_fn_attrib (fndecl)) + error_at (loc, "%<#pragma acc routine%> already applied to %D", fndecl); + + if (TREE_USED (fndecl) || (!is_defn && DECL_SAVED_TREE (fndecl))) + error_at (OMP_CLAUSE_LOCATION (TREE_PURPOSE (clauses)), + "%<#pragma acc routine%> must be applied before %s", + TREE_USED (fndecl) ? "use" : "definition"); + + /* Process for function attrib */ + tree dims = build_oacc_routine_dims (TREE_VALUE (clauses)); + replace_oacc_fn_attrib (fndecl, dims); + + /* Also attach as a declare. */ + DECL_ATTRIBUTES (fndecl) + = tree_cons (get_identifier ("omp declare target"), + clauses, DECL_ATTRIBUTES (fndecl)); +} + +/* Parse the OpenACC routine pragma. This has an optional '( name )' + component, which must resolve to a declared namespace-scope + function. The clauses are either processed directly (for a named + function), or defered until the immediatley following declaration + is parsed. +*/ + static void cp_parser_oacc_routine (cp_parser *parser, cp_token *pragma_tok, enum pragma_context context) { - tree name = NULL_TREE; - - //cp_lexer_consume_token (parser->lexer); + tree decl = NULL_TREE; + /* Create a dummy claue, to record location. */ + tree c_head = build_omp_clause (pragma_tok->location, OMP_CLAUSE_SEQ); - /* Scan for optional '( name )'. */ - if (cp_lexer_next_token_is (parser->lexer, CPP_OPEN_PAREN)) + if (context != pragma_external) + cp_parser_error (parser, "%<#pragma acc routine%> not at file scope"); + + /* Look for optional '( name )'. */ + if (cp_lexer_next_token_is (parser->lexer,CPP_OPEN_PAREN)) { cp_lexer_consume_token (parser->lexer); - name = cp_parser_id_expression (parser, /*template_p=*/false, - /*check_dependency_p=*/true, - /*template_p=*/NULL, - /*declarator_p=*/false, - /*optional_p=*/false); - if (name == error_mark_node) - return; + cp_token *token = cp_lexer_peek_token (parser->lexer); - if (cp_lexer_next_token_is_not (parser->lexer, CPP_CLOSE_PAREN)) + /* 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); + decl = cp_parser_lookup_name_simple (parser, id, token->location); + + if (decl == error_mark_node + || !cp_parser_require (parser, CPP_CLOSE_PAREN, RT_CLOSE_PAREN)) { - error_at (cp_lexer_peek_token (parser->lexer)->location, - "expected %<)%>"); + cp_parser_skip_to_pragma_eol (parser, pragma_tok); return; } - cp_lexer_consume_token (parser->lexer); - } - - /* If this routine construct doesn't explicitly have an optional 'name', - then handle it the same way as an omp declare simd. */ - if (!name) - { - cp_parser_omp_declare_simd (parser, pragma_tok, context, false); - cp_ensure_no_omp_declare_simd (parser); - return; } /* Build a chain of clauses. */ @@ -34437,64 +34416,27 @@ cp_parser_oacc_routine (cp_parser *parse cp_lexer_peek_token (parser->lexer), OACC_ROUTINE_CLAUSE_DEVICE_TYPE_MASK); - TREE_CHAIN (name) = clauses; - vec_safe_push (parser->named_oacc_routines, name); + /* Force clauses to be non-null, by attaching context to it. */ + clauses = tree_cons (c_head, clauses, NULL_TREE); + + if (decl) + cp_parser_finish_oacc_routine (parser, decl, clauses, true, false); + else + parser->oacc_routine = clauses; } -/* Finalize #pragma acc routine clauses after direct declarator has - been parsed, and put that into "omp declare target" attribute. */ +/* Apply any saved OpenACC routine clauses to a just-parsed + declaration. */ -static tree -cp_parser_late_parsing_oacc_routine (cp_parser *parser, tree attrs) +static void +cp_finalize_oacc_routine (cp_parser *parser, tree fndecl, bool is_defn) { - struct cp_token_cache *ce; - cp_omp_declare_simd_data *data = parser->oacc_routine; - int i; - - if (!data->error_seen && data->fndecl_seen) - { - error ("%<#pragma acc routine%> not immediately followed by " - "a single function declaration or definition"); - data->error_seen = true; - return attrs; - } - if (data->error_seen) - return attrs; - - tree c, cl = NULL_TREE; - - FOR_EACH_VEC_ELT (data->tokens, i, ce) + if (parser->oacc_routine) { - cp_parser_push_lexer_for_tokens (parser, ce); - parser->lexer->in_pragma = true; - gcc_assert (cp_lexer_peek_token (parser->lexer)->type == CPP_PRAGMA); - cp_token *pragma_tok = cp_lexer_consume_token (parser->lexer); - c = cp_parser_oacc_all_clauses (parser, OACC_ROUTINE_CLAUSE_MASK, - "#pragma acc routine", pragma_tok); - cp_parser_pop_lexer (parser); - - if (cl == NULL_TREE) - cl = c; - else if (c != NULL_TREE) - { - OMP_CLAUSE_CHAIN (c) = cl; - cl = c; - TREE_CHAIN (c) = attrs; - if (processing_template_decl) - ATTR_IS_DEPENDENT (c) = 1; - attrs = c; - } + cp_parser_finish_oacc_routine (parser, fndecl, parser->oacc_routine, + false, is_defn); + parser->oacc_routine = NULL_TREE; } - - tree dims = build_oacc_routine_dims (cl); - attrs = tree_cons (get_identifier ("oacc function"), dims, attrs); - - if (cl != NULL_TREE) - cl = tree_cons (NULL_TREE, cl, NULL_TREE); - - attrs = tree_cons (get_identifier ("omp declare target"), cl, attrs); - data->fndecl_seen = true; - return attrs; } /* Main entry point to OpenMP statement pragmas. */ @@ -34979,6 +34921,7 @@ cp_parser_pragma (cp_parser *parser, enu id = pragma_tok->pragma_kind; if (id != PRAGMA_OMP_DECLARE_REDUCTION && id != PRAGMA_OACC_ROUTINE) cp_ensure_no_omp_declare_simd (parser); + cp_ensure_no_oacc_routine (parser); switch (id) { case PRAGMA_GCC_PCH_PREPROCESS: Index: gcc/cp/parser.h =================================================================== --- gcc/cp/parser.h (revision 226912) +++ gcc/cp/parser.h (working copy) @@ -373,10 +373,9 @@ typedef struct GTY(()) cp_parser { necessary. */ cp_omp_declare_simd_data * GTY((skip)) cilk_simd_fn_info; - /* OpenACC specific parser information. */ - cp_omp_declare_simd_data * GTY((skip)) oacc_routine; - vec *named_oacc_routines; - + /* OpenACC routine clauses for subsequent decl/defn. */ + tree oacc_routine; + /* Nonzero if parsing a parameter list where 'auto' should trigger an implicit template parameter. */ bool auto_is_implicit_function_template_parm_p; Index: gcc/testsuite/c-c++-common/goacc/routine-2.c =================================================================== --- gcc/testsuite/c-c++-common/goacc/routine-2.c (revision 226912) +++ gcc/testsuite/c-c++-common/goacc/routine-2.c (working copy) @@ -1,6 +1,8 @@ void *malloc (__SIZE_TYPE__); void free (void *); +int fact (int n); + #pragma acc routine (fact) int Index: libgomp/testsuite/libgomp.oacc-c-c++-common/routine-2.c =================================================================== --- libgomp/testsuite/libgomp.oacc-c-c++-common/routine-2.c (revision 226912) +++ libgomp/testsuite/libgomp.oacc-c-c++-common/routine-2.c (working copy) @@ -6,8 +6,9 @@ #include #include -#pragma acc routine (fact) +int fact (int); +#pragma acc routine (fact) int fact (int n) {