2015-08-17  Nathan Sidwell  <nathan@codesourcery.com>

	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<c_token>);
-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 <tree, va_gc> *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 <stdio.h>
 #include <stdlib.h>
 
-#pragma acc routine (fact)
+int fact (int);
 
+#pragma acc routine (fact)
 
 int fact (int n)
 {
