diff mbox

[gomp4] Make front ends emit warnings when OpenACC routines directives lack parallelism

Message ID 599e8ecc-f9b8-e56f-b91f-358ee462377e@codesourcery.com
State New
Headers show

Commit Message

Cesar Philippidis May 2, 2017, 1:40 a.m. UTC
The language regarding OpenACC routines have changed in OpenACC 2.5 such
that the end user must explicitly specify one of gang, worker, vector or
seq partitioning. I guess some other compiler need those directives to
generate specialized versions of those functions accordingly. However,
GCC currently falls back to 'seq' partitioning, and that seems to be a
reasonable compromise in most cases.

This patch teaches the FE's how to emit a warning whenever a routine
directive doesn't include a partitioning clause. The rationale for
making this a warning vs an error is that it enables us to have some
backwards compatibility with OpenACC 2.0a.

Unfortunately, the fortran FE doesn't make use of
verify_oacc_routine_clauses because it parses/matches routines much
earlier than c/c++, so that part of the patch is slightly more involved.
I've also had to update a lot of test cases to make them conform to the
new routine behavior.

This patch has been committed to gomp-4_0-branch.

Cesar
diff mbox

Patch

2017-05-01  Cesar Philippidis  <cesar@codesourcery.com>

	gcc/fortran/
	* gfortran.h (enum oacc_function): Add OACC_FUNCTION_AUTO.
	* openmp.c (gfc_oacc_routine_dims): Return OACC_FUNCTION_AUTO when no
	parallelism was detected.
	(gfc_match_oacc_routine): Emit a warning when the user doesn't
	supply a gang, worker, vector or seq clause to an OpenACC routine
	construct.

	gcc/
	* omp-low.c (verify_oacc_routine_clauses): Emit a warning when the
	user doesn't supply a gang, worker, vector or seq clause to an
	OpenACC routine construct.

	gcc/testsuite/
	* c-c++-common/goacc-gomp/nesting-fail-1.c: Update usage and expected
	output of OpenACC routines.
	* c-c++-common/goacc/Wparentheses-1.c: Likewise.
	* c-c++-common/goacc/nesting-fail-1.c: Likewise.
	* c-c++-common/goacc/routine-1.c: Likewise.
	* c-c++-common/goacc/routine-2.c: Likewise.
	* c-c++-common/goacc/routine-5.c: Likewise.
	* c-c++-common/goacc/routine-level-of-parallelism-1.c: Likewise.
	* c-c++-common/goacc/routine-level-of-parallelism-2.c: Likewise.
	* c-c++-common/goacc/routine-nohost-1.c: Likewise.
	* c-c++-common/goacc/routine-nohost-2.c: Likewise.
	* g++.dg/goacc/routine-1.C: Likewise.
	* g++.dg/goacc/routine-2.C: Likewise.
	* g++.dg/goacc/template.C: Likewise.
	* gfortran.dg/goacc/dtype-1.f95: Likewise.
	* gfortran.dg/goacc/pr71704.f90: Likewise.
	* gfortran.dg/goacc/pr72741-2.f: Likewise.
	* gfortran.dg/goacc/routine-6.f90: Likewise.
	* gfortran.dg/goacc/routine-8.f90: Likewise.
	* gfortran.dg/goacc/routine-9.f90: Likewise.
	* gfortran.dg/goacc/routine-level-of-parallelism-1.f90: Likewise.
	* gfortran.dg/goacc/routine-without-clauses.f90: New test.

	libgomp/
	* testsuite/libgomp.oacc-c++/pr71959-a.C: Adjust test case conform
	to OpenACC 2.5 routines. 
	* testsuite/libgomp.oacc-c-c++-common/declare-2.c: Likewise.
	* testsuite/libgomp.oacc-c-c++-common/declare-3.c: Likewise.
	* testsuite/libgomp.oacc-c-c++-common/declare-4.c: Likewise.
	* testsuite/libgomp.oacc-c-c++-common/host_data-1.c: Likewise.
	* testsuite/libgomp.oacc-c-c++-common/loop-default.h: Likewise.
	* testsuite/libgomp.oacc-c-c++-common/mode-transitions.c: Likewise.
	* testsuite/libgomp.oacc-c-c++-common/parallel-loop-2.h: Likewise.
	* testsuite/libgomp.oacc-c-c++-common/routine-1.c: Likewise.
	* testsuite/libgomp.oacc-c-c++-common/routine-3.c: Likewise.
	* testsuite/libgomp.oacc-c-c++-common/routine-bind-nohost-1.c:
	Likewise.
	* testsuite/libgomp.oacc-fortran/host_data-2.f90: Likewise.
	* testsuite/libgomp.oacc-fortran/host_data-3.f: Likewise.
	* testsuite/libgomp.oacc-fortran/host_data-4.f90: Likewise.
	* testsuite/libgomp.oacc-fortran/routine-1.f90: Likewise.
	* testsuite/libgomp.oacc-fortran/routine-2.f90: Likewise.
	* testsuite/libgomp.oacc-fortran/routine-3.f90: Likewise.
	* testsuite/libgomp.oacc-fortran/routine-4.f90: Likewise.
	* testsuite/libgomp.oacc-fortran/routine-5.f90: Likewise.
	* testsuite/libgomp.oacc-fortran/routine-9.f90: Likewise.

diff --git a/gcc/fortran/gfortran.h b/gcc/fortran/gfortran.h
index 7913ac7..3c762a8 100644
--- a/gcc/fortran/gfortran.h
+++ b/gcc/fortran/gfortran.h
@@ -316,7 +316,8 @@  enum oacc_function
   OACC_FUNCTION_GANG,
   OACC_FUNCTION_WORKER,
   OACC_FUNCTION_VECTOR,
-  OACC_FUNCTION_SEQ
+  OACC_FUNCTION_SEQ,
+  OACC_FUNCTION_AUTO
 };
 
 /* Strings for all symbol attributes.  We use these for dumping the
diff --git a/gcc/fortran/openmp.c b/gcc/fortran/openmp.c
index 72c6669..88ccff2 100644
--- a/gcc/fortran/openmp.c
+++ b/gcc/fortran/openmp.c
@@ -2379,7 +2379,7 @@  static oacc_function
 gfc_oacc_routine_dims (gfc_omp_clauses *clauses)
 {
   int level = -1;
-  oacc_function ret = OACC_FUNCTION_SEQ;
+  oacc_function ret = OACC_FUNCTION_AUTO;
 
   if (clauses)
     {
@@ -2401,7 +2401,10 @@  gfc_oacc_routine_dims (gfc_omp_clauses *clauses)
 	  ret = OACC_FUNCTION_VECTOR;
 	}
       if (clauses->seq)
-	level = GOMP_DIM_MAX, mask |= GOMP_DIM_MASK (level);
+	{
+	  level = GOMP_DIM_MAX, mask |= GOMP_DIM_MASK (level);
+	  ret = OACC_FUNCTION_SEQ;
+	}
 
       if (mask != (mask & -mask))
 	ret = OACC_FUNCTION_NONE;
@@ -2505,6 +2508,12 @@  gfc_match_oacc_routine (void)
 	 know of any potential duplicate routine directives.  */
       seen_error = true;
     }
+  else if (dims == OACC_FUNCTION_AUTO)
+    {
+      gfc_warning (0, "Expected one of %<gang%>, %<worker%>, %<vector%> or "
+		   "%<seq%> clauses in !$ACC ROUTINE at %L", &old_loc);
+      dims = OACC_FUNCTION_SEQ;
+    }
 
   if (isym != NULL)
     {
diff --git a/gcc/omp-low.c b/gcc/omp-low.c
index cc209ba..88a1bb9 100644
--- a/gcc/omp-low.c
+++ b/gcc/omp-low.c
@@ -13239,8 +13239,10 @@  verify_oacc_routine_clauses (tree fndecl, tree *clauses, location_t loc,
       }
   if (c_level == NULL_TREE)
     {
-      /* OpenACC 2.5 makes this an error; for the current OpenACC 2.0a
-	 implementation add an implicit "seq" clause.  */
+      /* OpenACC 2.5 expects the user to supply one parallelism clause.  */
+      warning_at (loc, 0, "expecting one of %<gang%>, %<worker%>, %<vector%> "
+		  "or %<seq%> clauses");
+      inform (loc, "assigning %<seq%> parallelism to this routine");
       c_level = build_omp_clause (loc, OMP_CLAUSE_SEQ);
       OMP_CLAUSE_CHAIN (c_level) = *clauses;
       *clauses = c_level;
diff --git a/gcc/testsuite/c-c++-common/goacc-gomp/nesting-fail-1.c b/gcc/testsuite/c-c++-common/goacc-gomp/nesting-fail-1.c
index 1a33242..57eaa02 100644
--- a/gcc/testsuite/c-c++-common/goacc-gomp/nesting-fail-1.c
+++ b/gcc/testsuite/c-c++-common/goacc-gomp/nesting-fail-1.c
@@ -362,7 +362,7 @@  f_acc_data (void)
   }
 }
 
-#pragma acc routine
+#pragma acc routine seq
 void
 f_acc_loop (void)
 {
@@ -436,7 +436,7 @@  f_acc_loop (void)
     }
 }
 
-#pragma acc routine
+#pragma acc routine seq
 void
 f_acc_routine (void)
 {
diff --git a/gcc/testsuite/c-c++-common/goacc/Wparentheses-1.c b/gcc/testsuite/c-c++-common/goacc/Wparentheses-1.c
index 08265b6..9d35ec5 100644
--- a/gcc/testsuite/c-c++-common/goacc/Wparentheses-1.c
+++ b/gcc/testsuite/c-c++-common/goacc/Wparentheses-1.c
@@ -4,9 +4,9 @@ 
 int a, b, c;
 void bar (void);
 void baz (void);
-#pragma acc routine
+#pragma acc routine seq
 void bar2 (void);
-#pragma acc routine
+#pragma acc routine seq
 void baz2 (void);
 
 void
diff --git a/gcc/testsuite/c-c++-common/goacc/nesting-fail-1.c b/gcc/testsuite/c-c++-common/goacc/nesting-fail-1.c
index 93a9111..1462095 100644
--- a/gcc/testsuite/c-c++-common/goacc/nesting-fail-1.c
+++ b/gcc/testsuite/c-c++-common/goacc/nesting-fail-1.c
@@ -57,7 +57,7 @@  f_acc_data (void)
   }
 }
 
-#pragma acc routine
+#pragma acc routine seq
 void
 f_acc_routine (void)
 {
diff --git a/gcc/testsuite/c-c++-common/goacc/routine-1.c b/gcc/testsuite/c-c++-common/goacc/routine-1.c
index 0b56661..a4ecfd3 100644
--- a/gcc/testsuite/c-c++-common/goacc/routine-1.c
+++ b/gcc/testsuite/c-c++-common/goacc/routine-1.c
@@ -21,15 +21,15 @@  void seq (void)
 }
 
 
-#pragma acc routine
+#pragma acc routine /* { dg-warning "expecting one of" } */
 void bind_f_1 (void)
 {
 }
 
-#pragma acc routine
+#pragma acc routine /* { dg-warning "expecting one of" } */
 extern void bind_f_1 (void);
 
-#pragma acc routine (bind_f_1)
+#pragma acc routine (bind_f_1)/* { dg-warning "expecting one of" } */
 
 
 #pragma acc routine \
@@ -106,3 +106,16 @@  int main ()
 
   return 0;
 }
+
+/* { dg-warning "expecting one of" "" { target *-*-* } 35 } */
+/* { dg-warning "expecting one of" "" { target *-*-* } 41 } */
+/* { dg-warning "expecting one of" "" { target *-*-* } 45 } */
+/* { dg-warning "expecting one of" "" { target *-*-* } 50 } */
+/* { dg-warning "expecting one of" "" { target *-*-* } 56 } */
+/* { dg-warning "expecting one of" "" { target *-*-* } 60 } */
+/* { dg-warning "expecting one of" "" { target *-*-* } 64 } */
+/* { dg-warning "expecting one of" "" { target *-*-* } 70 } */
+/* { dg-warning "expecting one of" "" { target *-*-* } 74 } */
+/* { dg-warning "expecting one of" "" { target *-*-* } 78 } */
+/* { dg-warning "expecting one of" "" { target *-*-* } 84 } */
+/* { dg-warning "expecting one of" "" { target *-*-* } 88 } */
diff --git a/gcc/testsuite/c-c++-common/goacc/routine-2.c b/gcc/testsuite/c-c++-common/goacc/routine-2.c
index debf6d7..65c7963 100644
--- a/gcc/testsuite/c-c++-common/goacc/routine-2.c
+++ b/gcc/testsuite/c-c++-common/goacc/routine-2.c
@@ -1,171 +1,171 @@ 
 /* Test invalid use of clauses with OpenACC routine.  */
 
 extern float F;
-#pragma acc routine bind (F) /* { dg-error ".F. does not refer to a function" } */
+#pragma acc routine seq bind (F) /* { dg-error ".F. does not refer to a function" } */
 extern void F_1 (void);
 
 typedef int T;
-#pragma acc routine bind (T) /* { dg-error ".T. does not refer to a function" } */
+#pragma acc routine seq bind (T) /* { dg-error ".T. does not refer to a function" } */
 extern void T_1 (void);
 
 #pragma acc routine (nothing) gang /* { dg-error ".nothing. has not been declared" } */
 
-#pragma acc routine bind (bind_0) /* { dg-error ".bind_0. has not been declared" }*/
+#pragma acc routine seq bind (bind_0) /* { dg-error ".bind_0. has not been declared" }*/
 extern void bind_0 (void);
 
 extern void a(void), b(void);
 
-#pragma acc routine bind(a) bind(b) /* { dg-error "too many .bind. clauses" } */
+#pragma acc routine seq bind(a) bind(b) /* { dg-error "too many .bind. clauses" } */
 extern void bind_1 (void);
 
-#pragma acc routine bind(a) bind("b") /* { dg-error "too many .bind. clauses" } */
+#pragma acc routine seq bind(a) bind("b") /* { dg-error "too many .bind. clauses" } */
 extern void bind_2 (void);
 
-#pragma acc routine bind("a") bind("b") /* { dg-error "too many .bind. clauses" } */
+#pragma acc routine seq bind("a") bind("b") /* { dg-error "too many .bind. clauses" } */
 extern void bind_3 (void);
 
-#pragma acc routine nohost nohost /* { dg-error "too many .nohost. clauses" } */
+#pragma acc routine seq nohost nohost /* { dg-error "too many .nohost. clauses" } */
 extern void nohost (void);
 
 
 /* bind clause on first OpenACC routine directive but not on following.  */
 
 extern void a_bind_f_1 (void);
-#pragma acc routine (a_bind_f_1)
+#pragma acc routine (a_bind_f_1) seq
 
 
 #pragma acc routine \
-  bind (a_bind_f_1)
+  bind (a_bind_f_1) seq
 void a_bind_f_1_1 (void)
 {
 }
 
-#pragma acc routine /* { dg-error "missing .bind. clause when applying .#pragma acc routine. to .\[void \]*a_bind_f_1_1\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
+#pragma acc routine seq /* { dg-error "missing .bind. clause when applying .#pragma acc routine. to .\[void \]*a_bind_f_1_1\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
 extern void a_bind_f_1_1 (void);
 
-#pragma acc routine (a_bind_f_1_1) /* { dg-error "missing .bind. clause when applying .#pragma acc routine. to .\[void \]*a_bind_f_1_1\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
+#pragma acc routine (a_bind_f_1_1) seq /* { dg-error "missing .bind. clause when applying .#pragma acc routine. to .\[void \]*a_bind_f_1_1\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
 
 
 /* Non-sensical bind clause, but permitted.  */
 #pragma acc routine \
-  bind ("a_bind_f_2")
+  bind ("a_bind_f_2") seq
 void a_bind_f_2 (void)
 {
 }
 
-#pragma acc routine /* { dg-error "missing .bind. clause when applying .#pragma acc routine. to .\[void \]*a_bind_f_2\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
+#pragma acc routine seq /* { dg-error "missing .bind. clause when applying .#pragma acc routine. to .\[void \]*a_bind_f_2\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
 extern void a_bind_f_2 (void);
 
-#pragma acc routine (a_bind_f_2) /* { dg-error "missing .bind. clause when applying .#pragma acc routine. to .\[void \]*a_bind_f_2\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
+#pragma acc routine (a_bind_f_2) seq /* { dg-error "missing .bind. clause when applying .#pragma acc routine. to .\[void \]*a_bind_f_2\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
 
 
 #pragma acc routine \
-  bind ("a_bind_f_2")
+  bind ("a_bind_f_2") seq
 void a_bind_f_2_1 (void)
 {
 }
 
-#pragma acc routine /* { dg-error "missing .bind. clause when applying .#pragma acc routine. to .\[void \]*a_bind_f_2_1\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
+#pragma acc routine seq /* { dg-error "missing .bind. clause when applying .#pragma acc routine. to .\[void \]*a_bind_f_2_1\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
 extern void a_bind_f_2_1 (void);
 
-#pragma acc routine (a_bind_f_2_1) /* { dg-error "missing .bind. clause when applying .#pragma acc routine. to .\[void \]*a_bind_f_2_1\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
+#pragma acc routine (a_bind_f_2_1) seq /* { dg-error "missing .bind. clause when applying .#pragma acc routine. to .\[void \]*a_bind_f_2_1\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
 
 
 /* No bind clause on first OpenACC routine directive, but on following.  */
 
-#pragma acc routine
+#pragma acc routine seq
 extern void b_bind_f_1 (void);
 
 
-#pragma acc routine
+#pragma acc routine seq
 void b_bind_f_1_1 (void)
 {
 }
 
 #pragma acc routine \
-  bind (b_bind_f_1) /* { dg-error "incompatible .bind. clause when applying .#pragma acc routine. to .\[void \]*b_bind_f_1_1\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
+  bind (b_bind_f_1) seq /* { dg-error "incompatible .bind. clause when applying .#pragma acc routine. to .\[void \]*b_bind_f_1_1\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
 extern void b_bind_f_1_1 (void);
 
 #pragma acc routine (b_bind_f_1_1) \
-  bind (b_bind_f_1) /* { dg-error "incompatible .bind. clause when applying .#pragma acc routine. to .\[void \]*b_bind_f_1_1\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
+  bind (b_bind_f_1) seq /* { dg-error "incompatible .bind. clause when applying .#pragma acc routine. to .\[void \]*b_bind_f_1_1\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
 
 
 /* Non-sensical bind clause, but permitted.  */
-#pragma acc routine
+#pragma acc routine seq
 void b_bind_f_2 (void)
 {
 }
 
 #pragma acc routine \
-  bind ("b_bind_f_2") /* { dg-error "incompatible .bind. clause when applying .#pragma acc routine. to .\[void \]*b_bind_f_2\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
+  bind ("b_bind_f_2") seq /* { dg-error "incompatible .bind. clause when applying .#pragma acc routine. to .\[void \]*b_bind_f_2\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
 extern void b_bind_f_2 (void);
 
 #pragma acc routine (b_bind_f_2) \
-  bind ("b_bind_f_2") /* { dg-error "incompatible .bind. clause when applying .#pragma acc routine. to .\[void \]*b_bind_f_2\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
+  bind ("b_bind_f_2") seq /* { dg-error "incompatible .bind. clause when applying .#pragma acc routine. to .\[void \]*b_bind_f_2\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
 
 
-#pragma acc routine
+#pragma acc routine seq
 void b_bind_f_2_1 (void)
 {
 }
 
 #pragma acc routine \
-  bind ("b_bind_f_2") /* { dg-error "incompatible .bind. clause when applying .#pragma acc routine. to .\[void \]*b_bind_f_2_1\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
+  bind ("b_bind_f_2") seq /* { dg-error "incompatible .bind. clause when applying .#pragma acc routine. to .\[void \]*b_bind_f_2_1\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
 extern void b_bind_f_2_1 (void);
 
 #pragma acc routine (b_bind_f_2_1) \
-  bind ("b_bind_f_2") /* { dg-error "incompatible .bind. clause when applying .#pragma acc routine. to .\[void \]*b_bind_f_2_1\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
+  bind ("b_bind_f_2") seq /* { dg-error "incompatible .bind. clause when applying .#pragma acc routine. to .\[void \]*b_bind_f_2_1\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
 
 
 /* Non-matching bind clauses.  */
 
-#pragma acc routine
+#pragma acc routine seq
 void c_bind_f_1a (void)
 {
 }
 
-#pragma acc routine
+#pragma acc routine seq
 extern void c_bind_f_1b (void);
 
 
 #pragma acc routine \
-  bind (c_bind_f_1a)
+  bind (c_bind_f_1a) seq
 void c_bind_f_1_1 (void)
 {
 }
 
 #pragma acc routine \
-  bind (c_bind_f_1b) /* { dg-error "incompatible .bind. clause when applying .#pragma acc routine. to .\[void \]*c_bind_f_1_1\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
+  bind (c_bind_f_1b) seq /* { dg-error "incompatible .bind. clause when applying .#pragma acc routine. to .\[void \]*c_bind_f_1_1\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
 extern void c_bind_f_1_1 (void);
 
 #pragma acc routine (c_bind_f_1_1) \
-  bind (c_bind_f_1b) /* { dg-error "incompatible .bind. clause when applying .#pragma acc routine. to .\[void \]*c_bind_f_1_1\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
+  bind (c_bind_f_1b) seq /* { dg-error "incompatible .bind. clause when applying .#pragma acc routine. to .\[void \]*c_bind_f_1_1\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
 
 
 /* Non-sensical bind clause, but permitted.  */
 #pragma acc routine \
-  bind ("c_bind_f_2")
+  bind ("c_bind_f_2") seq
 void c_bind_f_2 (void)
 {
 }
 
 #pragma acc routine \
-  bind ("C_BIND_F_2") /* { dg-error "incompatible .bind. clause when applying .#pragma acc routine. to .\[void \]*c_bind_f_2\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
+  bind ("C_BIND_F_2") seq /* { dg-error "incompatible .bind. clause when applying .#pragma acc routine. to .\[void \]*c_bind_f_2\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
 extern void c_bind_f_2 (void);
 
 #pragma acc routine (c_bind_f_2) \
-  bind ("C_BIND_F_2") /* { dg-error "incompatible .bind. clause when applying .#pragma acc routine. to .\[void \]*c_bind_f_2\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
+  bind ("C_BIND_F_2") seq /* { dg-error "incompatible .bind. clause when applying .#pragma acc routine. to .\[void \]*c_bind_f_2\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
 
 
 #pragma acc routine \
-  bind ("c_bind_f_2")
+  bind ("c_bind_f_2") seq
 void c_bind_f_2_1 (void)
 {
 }
 
 #pragma acc routine \
-  bind ("C_BIND_F_2") /* { dg-error "incompatible .bind. clause when applying .#pragma acc routine. to .\[void \]*c_bind_f_2_1\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
+  bind ("C_BIND_F_2") seq /* { dg-error "incompatible .bind. clause when applying .#pragma acc routine. to .\[void \]*c_bind_f_2_1\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
 extern void c_bind_f_2_1 (void);
 
 #pragma acc routine (c_bind_f_2_1) \
-  bind ("C_BIND_F_2") /* { dg-error "incompatible .bind. clause when applying .#pragma acc routine. to .\[void \]*c_bind_f_2_1\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
+  bind ("C_BIND_F_2") seq /* { dg-error "incompatible .bind. clause when applying .#pragma acc routine. to .\[void \]*c_bind_f_2_1\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
diff --git a/gcc/testsuite/c-c++-common/goacc/routine-5.c b/gcc/testsuite/c-c++-common/goacc/routine-5.c
index f10651d..b7e8402 100644
--- a/gcc/testsuite/c-c++-common/goacc/routine-5.c
+++ b/gcc/testsuite/c-c++-common/goacc/routine-5.c
@@ -4,11 +4,11 @@ 
 
 struct PC
 {
-#pragma acc routine /* { dg-error ".#pragma acc routine. must be at file scope" } */
+#pragma acc routine seq /* { 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
+#pragma acc routine seq
 	 /* { 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 } } */
@@ -18,26 +18,26 @@  void PC1( /* { dg-bogus "variable or field .PC1. declared void" "TODO" { xfail c
 void PC2()
 {
   if (0)
-#pragma acc routine /* { dg-error ".#pragma acc routine. must be at file scope" } */
+#pragma acc routine seq /* { 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" } */
+#pragma acc routine seq /* { 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 () seq /* { dg-error "expected (function name|unqualified-id) before .\\). token" } */
+#pragma acc routine (+) seq /* { dg-error "expected (function name|unqualified-id) before .\\+. token" } */
+#pragma acc routine (?) seq /* { dg-error "expected (function name|unqualified-id) before .\\?. token" } */
+#pragma acc routine (:) seq /* { dg-error "expected (function name|unqualified-id) before .:. token" } */
+#pragma acc routine (4) seq /* { 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" } */
+#pragma acc routine ("4") seq /* { 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" } */
@@ -49,84 +49,84 @@  extern void R2(void);
 /* "#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" } */
+#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
 int a;
 
-#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by a single function declaration or definition" } */
+#pragma acc routine seq /* { 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 ".#pragma acc routine. not immediately followed by function declaration or definition" } */
+#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
 int b, fn2 (void);
 
-#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
+#pragma acc routine seq /* { 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" } */
+#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by a single function declaration or definition" } */
 int fn3 (void), b2;
 
-#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
+#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
 typedef struct c c;
 
-#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
+#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
 struct d {} d;
 
-#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" } */
+#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
+#pragma acc routine seq /* { 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 seq /* { 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" } */
+#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
+#pragma acc routine seq /* { 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" } */
+#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
+#pragma acc routine seq /* { 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" } */
+#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
+#pragma acc routine seq /* { 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" } */
+#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
+#pragma acc routine seq /* { 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
+#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
+#pragma acc routine seq
 int fn4 (void);
 
 int fn5a (void);
 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)
+#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
+#pragma acc routine (fn5a) seq
+#pragma acc routine (fn5b) seq
 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" } */
+#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
+#pragma acc routine (fn6a) seq /* { dg-error ".fn6a. has not been declared" } */
+#pragma acc routine (fn6b) seq /* { dg-error ".fn6b. has not been declared" } */
 int fn6 (void);
 
 #ifdef __cplusplus
 
-#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" "" { target c++ } } */
+#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" "" { target c++ } } */
 namespace f {}
 
 namespace g {}
 
-#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" "" { target c++ } } */
+#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" "" { target c++ } } */
 using namespace g;
 
-#pragma acc routine (g) /* { dg-error ".g. does not refer to a function" "" { target c++ } } */
+#pragma acc routine (g) seq /* { dg-error ".g. does not refer to a function" "" { target c++ } } */
 
 #endif /* __cplusplus */
 
-#pragma acc routine (a) /* { dg-error ".a. does not refer to a function" } */
+#pragma acc routine (a) seq /* { dg-error ".a. does not refer to a function" } */
   
-#pragma acc routine (c) /* { dg-error ".c. does not refer to a function" } */
+#pragma acc routine (c) seq /* { dg-error ".c. does not refer to a function" } */
 
 
 /* Static assert.  */
@@ -143,24 +143,24 @@  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 *-*-* } } */
+#pragma acc routine (f_static_assert) seq /* { dg-error ".#pragma acc routine. already applied to .\[void \]*f_static_assert" "TODO" { xfail *-*-* } } */
 
 
 /* __extension__ usage.  */
 
-#pragma acc routine
+#pragma acc routine seq
 __extension__ extern void ex1();
 #pragma acc routine (ex1) worker /* { dg-error "has already been marked as an accelerator routine" } */
 
-#pragma acc routine
+#pragma acc routine seq
 __extension__ __extension__ __extension__ __extension__ __extension__ void ex2()
 {
 }
 #pragma acc routine (ex2) worker /* { dg-error "has already been marked as an accelerator routine" } */
 
-#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
+#pragma acc routine seq /* { 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 (ex3) seq /* { dg-error ".ex3. does not refer to a function" } */
 
 
 /* "#pragma acc routine" must be applied before.  */
@@ -172,11 +172,11 @@  void Foo ()
   Bar ();
 }
 
-#pragma acc routine (Bar) // { dg-error ".#pragma acc routine. must be applied before use" }
+#pragma acc routine (Bar) seq // { dg-error ".#pragma acc routine. must be applied before use" }
 
 #pragma acc routine (Foo) gang // { dg-error ".#pragma acc routine. must be applied before definition" }
 
-#pragma acc routine (Baz) // { dg-error "not been declared" }
+#pragma acc routine (Baz) seq // { dg-error "not been declared" }
 
 
 /* OpenACC declare.  */
@@ -185,7 +185,7 @@  int vb1;		/* { dg-error "directive for use" } */
 extern int vb2;		/* { dg-error "directive for use" } */
 static int vb3;		/* { dg-error "directive for use" } */
 
-#pragma acc routine
+#pragma acc routine seq
 int
 func1 (int a)
 {
@@ -196,7 +196,7 @@  func1 (int a)
   return vb3;
 }
 
-#pragma acc routine
+#pragma acc routine seq
 int
 func2 (int a)
 {
@@ -214,7 +214,7 @@  extern int vb6;			/* { dg-error "clause used in" } */
 static int vb7;			/* { dg-error "clause used in" } */
 #pragma acc declare link (vb7)
 
-#pragma acc routine
+#pragma acc routine seq
 int
 func3 (int a)
 {
@@ -231,7 +231,7 @@  extern int vb9;
 static int vb10;
 #pragma acc declare create (vb10)
 
-#pragma acc routine
+#pragma acc routine seq
 int
 func4 (int a)
 {
@@ -249,7 +249,7 @@  extern int vb12;
 extern int vb13;
 #pragma acc declare device_resident (vb13)
 
-#pragma acc routine
+#pragma acc routine seq
 int
 func5 (int a)
 {
@@ -260,7 +260,7 @@  func5 (int a)
   return vb13;
 }
 
-#pragma acc routine
+#pragma acc routine seq
 int
 func6 (int a)
 {
diff --git a/gcc/testsuite/c-c++-common/goacc/routine-level-of-parallelism-1.c b/gcc/testsuite/c-c++-common/goacc/routine-level-of-parallelism-1.c
index 8f45499..e32d89d 100644
--- a/gcc/testsuite/c-c++-common/goacc/routine-level-of-parallelism-1.c
+++ b/gcc/testsuite/c-c++-common/goacc/routine-level-of-parallelism-1.c
@@ -428,20 +428,20 @@  extern void s_7 (void);
 void g_8 (void)
 {
 }
-#pragma acc routine (g_8) /* { dg-error "incompatible .seq. clause when applying .#pragma acc routine. to .\[void \]*g_8\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
+#pragma acc routine (g_8) seq /* { dg-error "incompatible .seq. clause when applying .#pragma acc routine. to .\[void \]*g_8\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
 
 #pragma acc routine \
   worker
 extern void w_8 (void);
-#pragma acc routine (w_8) /* { dg-error "incompatible .seq. clause when applying .#pragma acc routine. to .\[void \]*w_8\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
+#pragma acc routine (w_8) seq /* { dg-error "incompatible .seq. clause when applying .#pragma acc routine. to .\[void \]*w_8\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
 
 #pragma acc routine \
   vector
 extern void v_8 (void);
-#pragma acc routine (v_8) /* { dg-error "incompatible .seq. clause when applying .#pragma acc routine. to .\[void \]*v_8\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
+#pragma acc routine (v_8) seq /* { dg-error "incompatible .seq. clause when applying .#pragma acc routine. to .\[void \]*v_8\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
 
 extern void s_8 (void);
-#pragma acc routine (s_8)
+#pragma acc routine (s_8) /* { dg-warning "expecting one of" } */
 #pragma acc routine (s_8) \
   vector /* { dg-error "incompatible .vector. clause when applying .#pragma acc routine. to .\[void \]*s_8\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
 #pragma acc routine (s_8) \
diff --git a/gcc/testsuite/c-c++-common/goacc/routine-level-of-parallelism-2.c b/gcc/testsuite/c-c++-common/goacc/routine-level-of-parallelism-2.c
index 1803997..7f26ef1 100644
--- a/gcc/testsuite/c-c++-common/goacc/routine-level-of-parallelism-2.c
+++ b/gcc/testsuite/c-c++-common/goacc/routine-level-of-parallelism-2.c
@@ -28,46 +28,46 @@  extern void v_1 (void);
 
 #pragma acc routine seq
 extern void s_1_1 (void);
-#pragma acc routine (s_1_1)
+#pragma acc routine (s_1_1) /* { dg-warning "expecting one of" } */
 #pragma acc routine (s_1_1) seq
-#pragma acc routine (s_1_1)
+#pragma acc routine (s_1_1) /* { dg-warning "expecting one of" } */
 #pragma acc routine (s_1_1) seq
 
-#pragma acc routine
+#pragma acc routine  /* { dg-warning "expecting one of" } */
 extern void s_1_2 (void);
-#pragma acc routine (s_1_2)
+#pragma acc routine (s_1_2) /* { dg-warning "expecting one of" } */
 #pragma acc routine (s_1_2) seq
-#pragma acc routine (s_1_2)
+#pragma acc routine (s_1_2) /* { dg-warning "expecting one of" } */
 #pragma acc routine (s_1_2) seq
 
 extern void s_2_1 (void);
 #pragma acc routine (s_2_1) seq
-#pragma acc routine (s_2_1)
+#pragma acc routine (s_2_1) /* { dg-warning "expecting one of" } */
 #pragma acc routine (s_2_1) seq
-#pragma acc routine (s_2_1)
+#pragma acc routine (s_2_1) /* { dg-warning "expecting one of" } */
 #pragma acc routine (s_2_1) seq
 
 extern void s_2_2 (void);
-#pragma acc routine (s_2_2)
-#pragma acc routine (s_2_2)
+#pragma acc routine (s_2_2) /* { dg-warning "expecting one of" } */
+#pragma acc routine (s_2_2) /* { dg-warning "expecting one of" } */
 #pragma acc routine (s_2_2) seq
-#pragma acc routine (s_2_2)
+#pragma acc routine (s_2_2) /* { dg-warning "expecting one of" } */
 #pragma acc routine (s_2_2) seq
 
 #pragma acc routine seq
 void s_3_1 (void)
 {
 }
-#pragma acc routine (s_3_1)
+#pragma acc routine (s_3_1) /* { dg-warning "expecting one of" } */
 #pragma acc routine (s_3_1) seq
-#pragma acc routine (s_3_1)
+#pragma acc routine (s_3_1) /* { dg-warning "expecting one of" } */
 #pragma acc routine (s_3_1) seq
 
-#pragma acc routine
+#pragma acc routine seq
 void s_3_2 (void)
 {
 }
-#pragma acc routine (s_3_2)
+#pragma acc routine (s_3_2) /* { dg-warning "expecting one of" } */
 #pragma acc routine (s_3_2) seq
-#pragma acc routine (s_3_2)
+#pragma acc routine (s_3_2) /* { dg-warning "expecting one of" } */
 #pragma acc routine (s_3_2) seq
diff --git a/gcc/testsuite/c-c++-common/goacc/routine-nohost-1.c b/gcc/testsuite/c-c++-common/goacc/routine-nohost-1.c
index 17d6b03..800bcf6 100644
--- a/gcc/testsuite/c-c++-common/goacc/routine-nohost-1.c
+++ b/gcc/testsuite/c-c++-common/goacc/routine-nohost-1.c
@@ -3,46 +3,46 @@ 
 
 /* { dg-additional-options "-fdump-tree-oaccdevlow" } */
 
-#pragma acc routine nohost
+#pragma acc routine nohost seq
 int THREE(void)
 {
   return 3;
 }
 
-#pragma acc routine (THREE) nohost
+#pragma acc routine (THREE) nohost seq
 
-#pragma acc routine nohost
+#pragma acc routine nohost seq
 extern int THREE(void);
 
 
-#pragma acc routine nohost
+#pragma acc routine nohost seq
 extern void NOTHING(void);
 
-#pragma acc routine (NOTHING) nohost
+#pragma acc routine (NOTHING) nohost seq
 
 void NOTHING(void)
 {
 }
 
-#pragma acc routine nohost
+#pragma acc routine nohost seq
 extern void NOTHING(void);
 
-#pragma acc routine (NOTHING) nohost
+#pragma acc routine (NOTHING) nohost seq
 
 
 extern float ADD(float, float);
 
-#pragma acc routine (ADD) nohost
+#pragma acc routine (ADD) nohost seq
 
 float ADD(float x, float y)
 {
   return x + y;
 }
 
-#pragma acc routine nohost
+#pragma acc routine nohost seq
 extern float ADD(float, float);
 
-#pragma acc routine (ADD) nohost
+#pragma acc routine (ADD) nohost seq
 
 
 /* { dg-final { scan-tree-dump-times "Discarding function" 3 "oaccdevlow" } } */
diff --git a/gcc/testsuite/c-c++-common/goacc/routine-nohost-2.c b/gcc/testsuite/c-c++-common/goacc/routine-nohost-2.c
index 7402bfc..f172c38 100644
--- a/gcc/testsuite/c-c++-common/goacc/routine-nohost-2.c
+++ b/gcc/testsuite/c-c++-common/goacc/routine-nohost-2.c
@@ -1,41 +1,41 @@ 
 /* Test invalid usage of the nohost clause for OpenACC routine directive.
    Exercising different variants for declaring routines.  */
 
-#pragma acc routine
+#pragma acc routine seq
 int THREE_1(void)
 {
   return 3;
 }
 
 #pragma acc routine (THREE_1) \
-  nohost /* { dg-error "incompatible .nohost. clause when applying .#pragma acc routine. to .\[int \]*THREE_1\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
+  nohost seq /* { dg-error "incompatible .nohost. clause when applying .#pragma acc routine. to .\[int \]*THREE_1\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
 
 #pragma acc routine \
-  nohost /* { dg-error "incompatible .nohost. clause when applying .#pragma acc routine. to .\[int \]*THREE_1\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
+  nohost seq /* { dg-error "incompatible .nohost. clause when applying .#pragma acc routine. to .\[int \]*THREE_1\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
 extern int THREE_1(void);
 
 
-#pragma acc routine
+#pragma acc routine seq
 extern void NOTHING_1(void);
 
 #pragma acc routine (NOTHING_1) \
-  nohost /* { dg-error "incompatible .nohost. clause when applying .#pragma acc routine. to .\[void \]*NOTHING_1\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
+  nohost seq /* { dg-error "incompatible .nohost. clause when applying .#pragma acc routine. to .\[void \]*NOTHING_1\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
 
 void NOTHING_1(void)
 {
 }
 
 #pragma acc routine \
-  nohost /* { dg-error "incompatible .nohost. clause when applying .#pragma acc routine. to .\[void \]*NOTHING_1\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
+  nohost seq /* { dg-error "incompatible .nohost. clause when applying .#pragma acc routine. to .\[void \]*NOTHING_1\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
 extern void NOTHING_1(void);
 
 #pragma acc routine (NOTHING_1) \
-  nohost /* { dg-error "incompatible .nohost. clause when applying .#pragma acc routine. to .\[void \]*NOTHING_1\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
+  nohost seq /* { dg-error "incompatible .nohost. clause when applying .#pragma acc routine. to .\[void \]*NOTHING_1\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
 
 
 extern float ADD_1(float, float);
 
-#pragma acc routine (ADD_1)
+#pragma acc routine (ADD_1) seq
 
 float ADD_1(float x, float y)
 {
@@ -43,55 +43,55 @@  float ADD_1(float x, float y)
 }
 
 #pragma acc routine \
-  nohost /* { dg-error "incompatible .nohost. clause when applying .#pragma acc routine. to .\[float \]*ADD_1\[\\(float, \\)\]*., which has already been marked as an accelerator routine" } */
+  nohost seq /* { dg-error "incompatible .nohost. clause when applying .#pragma acc routine. to .\[float \]*ADD_1\[\\(float, \\)\]*., which has already been marked as an accelerator routine" } */
 extern float ADD_1(float, float);
 
 #pragma acc routine (ADD_1) \
-  nohost /* { dg-error "incompatible .nohost. clause when applying .#pragma acc routine. to .\[float \]*ADD_1\[\\(float, \\)\]*., which has already been marked as an accelerator routine" } */
+  nohost seq /* { dg-error "incompatible .nohost. clause when applying .#pragma acc routine. to .\[float \]*ADD_1\[\\(float, \\)\]*., which has already been marked as an accelerator routine" } */
 
 
 /* The same again, but with/without nohost reversed.  */
 
 #pragma acc routine \
-  nohost
+  nohost seq
 int THREE_2(void)
 {
   return 3;
 }
 
-#pragma acc routine (THREE_2) /* { dg-error "missing .nohost. clause when applying .#pragma acc routine. to .\[int \]*THREE_2\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
+#pragma acc routine (THREE_2) seq /* { dg-error "missing .nohost. clause when applying .#pragma acc routine. to .\[int \]*THREE_2\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
 
-#pragma acc routine /* { dg-error "missing .nohost. clause when applying .#pragma acc routine. to .\[int \]*THREE_2\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
+#pragma acc routine seq /* { dg-error "missing .nohost. clause when applying .#pragma acc routine. to .\[int \]*THREE_2\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
 extern int THREE_2(void);
 
 
 #pragma acc routine \
-  nohost
+  nohost seq
 extern void NOTHING_2(void);
 
-#pragma acc routine (NOTHING_2) /* { dg-error "missing .nohost. clause when applying .#pragma acc routine. to .\[void \]*NOTHING_2\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
+#pragma acc routine (NOTHING_2) seq /* { dg-error "missing .nohost. clause when applying .#pragma acc routine. to .\[void \]*NOTHING_2\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
 
 void NOTHING_2(void)
 {
 }
 
-#pragma acc routine /* { dg-error "missing .nohost. clause when applying .#pragma acc routine. to .\[void \]*NOTHING_2\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
+#pragma acc routine seq /* { dg-error "missing .nohost. clause when applying .#pragma acc routine. to .\[void \]*NOTHING_2\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
 extern void NOTHING_2(void);
 
-#pragma acc routine (NOTHING_2) /* { dg-error "missing .nohost. clause when applying .#pragma acc routine. to .\[void \]*NOTHING_2\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
+#pragma acc routine (NOTHING_2) seq /* { dg-error "missing .nohost. clause when applying .#pragma acc routine. to .\[void \]*NOTHING_2\[\\(\\)\]*., which has already been marked as an accelerator routine" } */
 
 
 extern float ADD_2(float, float);
 
 #pragma acc routine (ADD_2) \
-  nohost
+  nohost seq
 
 float ADD_2(float x, float y)
 {
   return x + y;
 }
 
-#pragma acc routine /* { dg-error "missing .nohost. clause when applying .#pragma acc routine. to .\[float \]*ADD_2\[\\(float, \\)\]*., which has already been marked as an accelerator routine" } */
+#pragma acc routine seq /* { dg-error "missing .nohost. clause when applying .#pragma acc routine. to .\[float \]*ADD_2\[\\(float, \\)\]*., which has already been marked as an accelerator routine" } */
 extern float ADD_2(float, float);
 
-#pragma acc routine (ADD_2) /* { dg-error "missing .nohost. clause when applying .#pragma acc routine. to .\[float \]*ADD_2\[\\(float, \\)\]*., which has already been marked as an accelerator routine" } */
+#pragma acc routine (ADD_2) seq /* { dg-error "missing .nohost. clause when applying .#pragma acc routine. to .\[float \]*ADD_2\[\\(float, \\)\]*., which has already been marked as an accelerator routine" } */
diff --git a/gcc/testsuite/g++.dg/goacc/routine-1.C b/gcc/testsuite/g++.dg/goacc/routine-1.C
index a73a73d..9257324 100644
--- a/gcc/testsuite/g++.dg/goacc/routine-1.C
+++ b/gcc/testsuite/g++.dg/goacc/routine-1.C
@@ -4,10 +4,10 @@  namespace N
 {
   extern void foo1();
   extern void foo2();
-#pragma acc routine (foo1)
-#pragma acc routine
+#pragma acc routine (foo1) seq
+#pragma acc routine seq
   void foo3()
   {
   }
 }
-#pragma acc routine (N::foo2)
+#pragma acc routine (N::foo2) seq
diff --git a/gcc/testsuite/g++.dg/goacc/routine-2.C b/gcc/testsuite/g++.dg/goacc/routine-2.C
index 939dfba..846f388 100644
--- a/gcc/testsuite/g++.dg/goacc/routine-2.C
+++ b/gcc/testsuite/g++.dg/goacc/routine-2.C
@@ -2,7 +2,7 @@ 
 
 template <typename T>
 extern T one_d();
-#pragma acc routine (one_d) nohost /* { dg-error "names a set of overloads" } */
+#pragma acc routine (one_d) nohost seq /* { dg-error "names a set of overloads" } */
 
 template <typename T>
 T
@@ -10,41 +10,41 @@  one()
 {
   return 1;
 }
-#pragma acc routine (one) bind(one_d) /* { dg-error "names a set of overloads" } */
+#pragma acc routine (one) bind(one_d) seq /* { dg-error "names a set of overloads" } */
 
 int incr (int);
 float incr (float);
 
-#pragma acc routine (incr) /* { dg-error "names a set of overloads" } */
+#pragma acc routine (incr) seq /* { dg-error "names a set of overloads" } */
 
 
 int sum (int, int);
 
 namespace foo {
-#pragma acc routine (sum)
+#pragma acc routine (sum) seq
   int sub (int, int);
 }
 
-#pragma acc routine (foo::sub)
+#pragma acc routine (foo::sub) seq
 
 /* It's strange to apply a routine directive to subset of overloaded
    functions, but that is permissible in OpenACC 2.x.  */
 
 int decr (int a);
 
-#pragma acc routine
+#pragma acc routine seq
 float decr (float a);
 
 /* Bind clause.  */
 
-#pragma acc routine
+#pragma acc routine seq
 float
 mult (float a, float b)
 {
   return a * b;
 }
 
-#pragma acc routine bind(mult) /* { dg-error "bind identifier 'mult' is not compatible with function 'broken_mult1'" } */
+#pragma acc routine bind(mult) seq /* { dg-error "bind identifier 'mult' is not compatible with function 'broken_mult1'" } */
 float
 broken_mult1 (int a, int b)
 {
@@ -52,51 +52,51 @@  broken_mult1 (int a, int b)
 }
 
 /* This should result in a link error, but it's valid for a compile test.  */
-#pragma acc routine bind("mult")
+#pragma acc routine bind("mult") seq
 float
 broken_mult2 (float a, float b)
 {
   return a + b;
 }
-#pragma acc routine (broken_mult2) bind("mult")
-#pragma acc routine bind("mult")
+#pragma acc routine (broken_mult2) bind("mult") seq
+#pragma acc routine bind("mult") seq
 extern float broken_mult2 (float, float);
 
-#pragma acc routine bind(sum2) /* { dg-error "'sum2' has not been declared" } */
+#pragma acc routine bind(sum2) seq /* { dg-error "'sum2' has not been declared" } */
 int broken_mult3 (int a, int b);
 
-#pragma acc routine bind(foo::sub)
+#pragma acc routine bind(foo::sub) seq
 int broken_mult4 (int a, int b);
-#pragma acc routine (broken_mult4) bind(foo::sub)
-#pragma acc routine bind(foo::sub)
+#pragma acc routine (broken_mult4) bind(foo::sub) seq
+#pragma acc routine bind(foo::sub) seq
 extern int broken_mult4 (int a, int b);
 
 namespace bar
 {
-#pragma acc routine bind (mult)
+#pragma acc routine bind (mult) seq
   float working_mult (float a, float b)
   {
     return a * b;
   }
-#pragma acc routine (working_mult) bind (mult)
-#pragma acc routine bind (mult)
+#pragma acc routine (working_mult) bind (mult) seq
+#pragma acc routine bind (mult) seq
   float working_mult (float, float);
 }
 
-#pragma acc routine
+#pragma acc routine seq
 int div (int, int);
 
-#pragma acc routine
+#pragma acc routine seq
 float div (float, float);
 
-#pragma acc routine bind (div) /* { dg-error "'div' names a set of overloads" } */
+#pragma acc routine bind (div) seq /* { dg-error "'div' names a set of overloads" } */
 float
 my_div (float a, float b)
 {
   return a / b;
 }
 
-#pragma acc routine bind (other_div) /* { dg-error "'other_div' has not been declared" } */
+#pragma acc routine bind (other_div) seq /* { dg-error "'other_div' has not been declared" } */
 float
 my_div2 (float a, float b)
 {
@@ -105,21 +105,21 @@  my_div2 (float a, float b)
 
 int div_var;
 
-#pragma acc routine bind (div_var) /* { dg-error "'div_var' does not refer to a function" } */
+#pragma acc routine bind (div_var) seq /* { dg-error "'div_var' does not refer to a function" } */
 float my_div3 (float, float);
 
-#pragma acc routine bind (div_var) /* { dg-error "'div_var' does not refer to a function" } */
+#pragma acc routine bind (div_var) seq /* { dg-error "'div_var' does not refer to a function" } */
 float my_div4 (float, float);
 
-#pragma acc routine bind (%) /* { dg-error "expected identifier or character string literal" } */
+#pragma acc routine bind (%) seq /* { dg-error "expected identifier or character string literal" } */
 float my_div5 (float, float);
 
-#pragma acc routine bind ("") /* { dg-error "bind argument must not be an empty string" } */
+#pragma acc routine bind ("") seq /* { dg-error "bind argument must not be an empty string" } */
 float my_div6 (float, float);
 
 struct astruct
 {
-  #pragma acc routine /* { dg-error ".#pragma acc routine. must be at file scope" } */
+  #pragma acc routine seq /* { dg-error ".#pragma acc routine. must be at file scope" } */
   int sum (int a, int b)
   {
     return a + b;
@@ -128,7 +128,7 @@  struct astruct
 
 class aclass
 {
-  #pragma acc routine /* { dg-error ".#pragma acc routine. must be at file scope" } */
+  #pragma acc routine seq /* { dg-error ".#pragma acc routine. must be at file scope" } */
   int sum (int a, int b)
   {
     return a + b;
diff --git a/gcc/testsuite/g++.dg/goacc/template.C b/gcc/testsuite/g++.dg/goacc/template.C
index 056398d..4bc2596 100644
--- a/gcc/testsuite/g++.dg/goacc/template.C
+++ b/gcc/testsuite/g++.dg/goacc/template.C
@@ -1,4 +1,4 @@ 
-#pragma acc routine
+#pragma acc routine seq
 template <typename T> T
 accDouble(int val)
 {
diff --git a/gcc/testsuite/gfortran.dg/goacc/dtype-1.f95 b/gcc/testsuite/gfortran.dg/goacc/dtype-1.f95
index 5919ae4..f24b60f 100644
--- a/gcc/testsuite/gfortran.dg/goacc/dtype-1.f95
+++ b/gcc/testsuite/gfortran.dg/goacc/dtype-1.f95
@@ -108,63 +108,63 @@  end program dtype
 !! ACC ROUTINE:
 
 subroutine sr1 ()
-  !$acc routine device_type (nvidia) gang
+  !$acc routine seq device_type (nvidia) gang
 end subroutine sr1
 
 subroutine sr2 ()
-  !$acc routine dtype (nvidia) worker
+  !$acc routine seq dtype (nvidia) worker
 end subroutine sr2
 
 subroutine sr3 ()
-  !$acc routine device_type (nvidia) vector
+  !$acc routine seq device_type (nvidia) vector
 end subroutine sr3
 
 subroutine sr4 ()
-  !$acc routine device_type (nvidia) seq
+  !$acc routine seq device_type (nvidia) seq
 end subroutine sr4
 
 subroutine sr5 ()
-  !$acc routine dtype (nvidia) bind (foo)
+  !$acc routine seq dtype (nvidia) bind (foo)
 end subroutine sr5
 
 subroutine sr1a ()
-  !$acc routine device_type (nvidia) gang device_type (*) seq
+  !$acc routine seq device_type (nvidia) gang device_type (*) seq
 end subroutine sr1a
 
 subroutine sr2a ()
-  !$acc routine dtype (nvidia) worker dtype (*) seq
+  !$acc routine seq dtype (nvidia) worker dtype (*) seq
 end subroutine sr2a
 
 subroutine sr3a ()
-  !$acc routine dtype (nvidia) vector device_type (*) seq
+  !$acc routine seq dtype (nvidia) vector device_type (*) seq
 end subroutine sr3a
 
 subroutine sr4a ()
-  !$acc routine device_type (nvidia) seq device_type (*) worker
+  !$acc routine seq device_type (nvidia) seq device_type (*) worker
 end subroutine sr4a
 
 subroutine sr5a ()
-  !$acc routine device_type (nvidia) bind (foo) dtype (*) seq
+  !$acc routine seq device_type (nvidia) bind (foo) dtype (*) seq
 end subroutine sr5a
 
 subroutine sr1b ()
-  !$acc routine dtype (gpu) gang dtype (*) seq
+  !$acc routine seq dtype (gpu) gang dtype (*) seq
 end subroutine sr1b
 
 subroutine sr2b ()
-  !$acc routine dtype (gpu) worker device_type (*) seq
+  !$acc routine seq dtype (gpu) worker device_type (*) seq
 end subroutine sr2b
 
 subroutine sr3b ()
-  !$acc routine device_type (gpu) vector device_type (*) seq
+  !$acc routine seq device_type (gpu) vector device_type (*) seq
 end subroutine sr3b
 
 subroutine sr4b ()
-  !$acc routine device_type (gpu) seq device_type (*) worker
+  !$acc routine seq device_type (gpu) seq device_type (*) worker
 end subroutine sr4b
 
 subroutine sr5b ()
-  !$acc routine dtype (gpu) bind (foo) device_type (*) seq
+  !$acc routine seq dtype (gpu) bind (foo) device_type (*) seq
 end subroutine sr5b
 
 ! { dg-final { scan-tree-dump-times "oacc_parallel device_type\\(nvidia\\) \\\[ async\\(1\\) wait\\(1\\) num_gangs\\(100\\) num_workers\\(100\\) vector_length\\(32\\) \\\]" 1 "omplower" } }
diff --git a/gcc/testsuite/gfortran.dg/goacc/pr71704.f90 b/gcc/testsuite/gfortran.dg/goacc/pr71704.f90
index 0235e85..92d0c71 100644
--- a/gcc/testsuite/gfortran.dg/goacc/pr71704.f90
+++ b/gcc/testsuite/gfortran.dg/goacc/pr71704.f90
@@ -2,7 +2,7 @@ 
 ! { dg-do compile }
 
 real function f1 ()
-!$acc routine (f1)
+!$acc routine (f1) seq
   f1 = 1
 end
 
diff --git a/gcc/testsuite/gfortran.dg/goacc/pr72741-2.f b/gcc/testsuite/gfortran.dg/goacc/pr72741-2.f
index 5865144..0f0f7df 100644
--- a/gcc/testsuite/gfortran.dg/goacc/pr72741-2.f
+++ b/gcc/testsuite/gfortran.dg/goacc/pr72741-2.f
@@ -1,5 +1,5 @@ 
       SUBROUTINE v_1
-!$ACC ROUTINE
+!$ACC ROUTINE SEQ
 !$ACC ROUTINE ! { dg-error "ACC ROUTINE already applied" }
 !$ACC ROUTINE GANG ! { dg-error "ACC ROUTINE already applied" }
 !$ACC ROUTINE SEQ ! { dg-error "ACC ROUTINE already applied" }
@@ -13,7 +13,7 @@ 
 !$ACC ROUTINE (g_1) GANG
 !$ACC ROUTINE (g_1) VECTOR ! { dg-error "ACC ROUTINE already applied" }
 !$ACC ROUTINE (g_1) SEQ ! { dg-error "ACC ROUTINE already applied" }
-!$ACC ROUTINE (g_1) ! { dg-error "ACC ROUTINE already applied" }
+!$ACC ROUTINE (g_1) WORKER! { dg-error "ACC ROUTINE already applied" }
 !$ACC ROUTINE (g_1) ! { dg-error "ACC ROUTINE already applied" }
 
       CALL v_1
diff --git a/gcc/testsuite/gfortran.dg/goacc/routine-6.f90 b/gcc/testsuite/gfortran.dg/goacc/routine-6.f90
index ee43935..5c1f652 100644
--- a/gcc/testsuite/gfortran.dg/goacc/routine-6.f90
+++ b/gcc/testsuite/gfortran.dg/goacc/routine-6.f90
@@ -4,7 +4,7 @@  module m
 contains
   subroutine subr5 (x) 
   implicit none
-  !$acc routine (subr5)
+  !$acc routine (subr5) seq
   integer, intent(inout) :: x
   if (x < 1) then
      x = 1
@@ -18,7 +18,7 @@  program main
   implicit none
   interface
     function subr6 (x) 
-    !$acc routine (subr6) ! { dg-error "without list is allowed in interface" }
+    !$acc routine (subr6) seq ! { dg-error "without list is allowed in interface" }
     integer, intent (in) :: x
     integer :: subr6
     end function subr6
@@ -26,13 +26,13 @@  program main
   integer, parameter :: n = 10
   integer :: a(n), i
   external :: subr2
-  !$acc routine (subr2)
+  !$acc routine (subr2) seq
 
   external :: R1, R2
-  !$acc routine (R1 R2 R3) ! { dg-error "Syntax error in \\!\\\$ACC ROUTINE \\( NAME \\) at \\(1\\), expecting .\\). after NAME" }
-  !$acc routine (R1, R2, R3) ! { dg-error "Syntax error in \\!\\\$ACC ROUTINE \\( NAME \\) at \\(1\\), expecting .\\). after NAME" }
-  !$acc routine (R1)
-  !$acc routine (R2)
+  !$acc routine (R1 R2 R3) seq ! { dg-error "Syntax error in \\!\\\$ACC ROUTINE \\( NAME \\) at \\(1\\), expecting .\\). after NAME" }
+  !$acc routine (R1, R2, R3) seq ! { dg-error "Syntax error in \\!\\\$ACC ROUTINE \\( NAME \\) at \\(1\\), expecting .\\). after NAME" }
+  !$acc routine (R1) seq
+  !$acc routine (R2) seq
 
   !$acc parallel
   !$acc loop
@@ -44,7 +44,7 @@  program main
 end program main
 
 subroutine subr1 (x) 
-  !$acc routine
+  !$acc routine seq
   integer, intent(inout) :: x
   if (x < 1) then
      x = 1
@@ -63,7 +63,7 @@  subroutine subr2 (x)
 end subroutine subr2
 
 subroutine subr3 (x) 
-  !$acc routine (subr3)
+  !$acc routine (subr3) seq
   integer, intent(inout) :: x
   if (x < 1) then
      x = 1
@@ -73,7 +73,7 @@  subroutine subr3 (x)
 end subroutine subr3
 
 subroutine subr4 (x) 
-  !$acc routine (subr4)
+  !$acc routine (subr4) seq
   integer, intent(inout) :: x
   if (x < 1) then
      x = 1
diff --git a/gcc/testsuite/gfortran.dg/goacc/routine-8.f90 b/gcc/testsuite/gfortran.dg/goacc/routine-8.f90
index c903915..beca43f 100644
--- a/gcc/testsuite/gfortran.dg/goacc/routine-8.f90
+++ b/gcc/testsuite/gfortran.dg/goacc/routine-8.f90
@@ -4,7 +4,7 @@  program main
   interface
      function s_1 (a)
        integer a
-       !$acc routine
+       !$acc routine seq
      end function s_1
   end interface
 
@@ -18,7 +18,7 @@  program main
   interface
      function s_3 (a)
        integer a
-       !$acc routine (s_3) ! { dg-error "Only the ..ACC ROUTINE form without list is allowed in interface block" }
+       !$acc routine (s_3) seq ! { dg-error "Only the ..ACC ROUTINE form without list is allowed in interface block" }
      end function s_3
   end interface
 
diff --git a/gcc/testsuite/gfortran.dg/goacc/routine-9.f90 b/gcc/testsuite/gfortran.dg/goacc/routine-9.f90
index 590e594..4027471 100644
--- a/gcc/testsuite/gfortran.dg/goacc/routine-9.f90
+++ b/gcc/testsuite/gfortran.dg/goacc/routine-9.f90
@@ -7,9 +7,9 @@  contains
   subroutine subr5 (x)
     implicit none
     integer extfunc
-    !$acc routine (subr5)
-    !$acc routine (extfunc)
-    !$acc routine (m1int) ! { dg-error "invalid function name" }
+    !$acc routine (subr5) seq
+    !$acc routine (extfunc) seq
+    !$acc routine (m1int) seq ! { dg-error "invalid function name" }
     integer, intent(inout) :: x
     if (x < 1) then
        x = 1
@@ -29,13 +29,13 @@  program main
   end interface
   integer, parameter :: n = 10
   integer :: a(n), i
-  !$acc routine (subr1) ! { dg-error "invalid function name" }
+  !$acc routine (subr1) seq ! { dg-error "invalid function name" }
   external :: subr2
-  !$acc routine (subr2)
+  !$acc routine (subr2) seq
 
   external :: R1, R2
-  !$acc routine (R1)
-  !$acc routine (R2)
+  !$acc routine (R1) seq
+  !$acc routine (R2) seq
 
   !$acc parallel
   !$acc loop
@@ -47,7 +47,7 @@  program main
 end program main
 
 subroutine subr1 (x)
-  !$acc routine
+  !$acc routine seq
   integer, intent(inout) :: x
   if (x < 1) then
      x = 1
@@ -66,7 +66,7 @@  subroutine subr2 (x)
 end subroutine subr2
 
 subroutine subr3 (x)
-  !$acc routine (subr3)
+  !$acc routine (subr3) seq
   integer, intent(inout) :: x
   if (x < 1) then
      x = 1
@@ -76,7 +76,7 @@  subroutine subr3 (x)
 end subroutine subr3
 
 subroutine subr4 (x)
-  !$acc routine (subr4)
+  !$acc routine (subr4) seq
   integer, intent(inout) :: x
   if (x < 1) then
      x = 1
@@ -86,7 +86,7 @@  subroutine subr4 (x)
 end subroutine subr4
 
 subroutine subr10 (x)
-  !$acc routine (subr10) device ! { dg-error "Unclassifiable OpenACC directive" }
+  !$acc routine (subr10) seq device ! { dg-error "Unclassifiable OpenACC directive" }
   integer, intent(inout) :: x
   if (x < 1) then
      x = 1
diff --git a/gcc/testsuite/gfortran.dg/goacc/routine-level-of-parallelism-1.f90 b/gcc/testsuite/gfortran.dg/goacc/routine-level-of-parallelism-1.f90
index 364a058..7c245ce 100644
--- a/gcc/testsuite/gfortran.dg/goacc/routine-level-of-parallelism-1.f90
+++ b/gcc/testsuite/gfortran.dg/goacc/routine-level-of-parallelism-1.f90
@@ -9,7 +9,7 @@  subroutine g_1 ! { dg-warning "region is gang partitioned but does not contain g
 end subroutine g_1
 
 subroutine s_1_2a
-  !$acc routine
+  !$acc routine seq
 end subroutine s_1_2a
 
 subroutine s_1_2b
@@ -17,7 +17,7 @@  subroutine s_1_2b
 end subroutine s_1_2b
 
 subroutine s_1_2c
-  !$acc routine (s_1_2c)
+  !$acc routine (s_1_2c) seq
 end subroutine s_1_2c
 
 subroutine s_1_2d
@@ -27,7 +27,7 @@  end subroutine s_1_2d
 module s_2
 contains
   subroutine s_2_1a
-    !$acc routine
+    !$acc routine seq
   end subroutine s_2_1a
 
   subroutine s_2_1b
@@ -35,7 +35,7 @@  contains
   end subroutine s_2_1b
 
   subroutine s_2_1c
-    !$acc routine (s_2_1c)
+    !$acc routine (s_2_1c) seq
   end subroutine s_2_1c
 
   subroutine s_2_1d
@@ -50,7 +50,7 @@  subroutine test
   interface
      function s_3_1a (a)
        integer a
-       !$acc routine
+       !$acc routine seq
      end function s_3_1a
   end interface
 
diff --git a/gcc/testsuite/gfortran.dg/goacc/routine-without-clauses.f90 b/gcc/testsuite/gfortran.dg/goacc/routine-without-clauses.f90
new file mode 100644
index 0000000..570c8ed
--- /dev/null
+++ b/gcc/testsuite/gfortran.dg/goacc/routine-without-clauses.f90
@@ -0,0 +1,36 @@ 
+! Test the OpenACC routine directive when it has no gang, worker
+! vector, or seq partitioning clauses.
+
+! { dg-do compile }
+
+subroutine s1
+  !$acc routine ! { dg-warning "Expected one of" }
+end subroutine s1
+
+integer function f1 ()
+  !$acc routine ! { dg-warning "Expected one of" }
+end function f1
+
+module m
+contains
+  subroutine s2
+    !$acc routine ! { dg-warning "Expected one of" }
+  end subroutine s2
+
+  integer function f2 ()
+    !$acc routine ! { dg-warning "Expected one of" }
+  end function f2
+end module m
+
+program t
+  implicit none
+  interface
+     subroutine s3
+       !$acc routine ! { dg-warning "Expected one of" }
+     end subroutine s3
+
+     integer function f3 ()
+       !$acc routine ! { dg-warning "Expected one of" }
+     end function f3
+  end interface
+end program t
diff --git a/libgomp/testsuite/libgomp.oacc-c++/pr71959-a.C b/libgomp/testsuite/libgomp.oacc-c++/pr71959-a.C
index 9486512..82d6c22 100644
--- a/libgomp/testsuite/libgomp.oacc-c++/pr71959-a.C
+++ b/libgomp/testsuite/libgomp.oacc-c++/pr71959-a.C
@@ -8,13 +8,13 @@  struct Iter
   int *point () const asm("_ZNK4Iter5pointEv");
 };
 
-#pragma acc routine
+#pragma acc routine seq
 void  Iter::ctor (int *cursor_)
 {
   cursor = cursor_;
 }
 
-#pragma acc routine
+#pragma acc routine seq
 int *Iter::point () const
 {
   return cursor;
@@ -22,7 +22,7 @@  int *Iter::point () const
 
 void apply (int (*fn)(), Iter out) asm ("_ZN5Apply5applyEPFivE4Iter");
 
-#pragma acc routine
+#pragma acc routine seq
 void apply (int (*fn)(), struct Iter out)
 { *out.point() = fn (); }
 
diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/declare-2.c b/libgomp/testsuite/libgomp.oacc-c-c++-common/declare-2.c
index 2078a33..4f06ffd 100644
--- a/libgomp/testsuite/libgomp.oacc-c-c++-common/declare-2.c
+++ b/libgomp/testsuite/libgomp.oacc-c-c++-common/declare-2.c
@@ -7,7 +7,7 @@ 
 float c[N];
 #pragma acc declare device_resident (c)
 
-#pragma acc routine
+#pragma acc routine seq
 float
 subr2 (float a)
 {
@@ -25,7 +25,7 @@  subr2 (float a)
 float b[N];
 #pragma acc declare copyin (b)
 
-#pragma acc routine
+#pragma acc routine seq
 float
 subr1 (float a)
 {
diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/declare-3.c b/libgomp/testsuite/libgomp.oacc-c-c++-common/declare-3.c
index c3a2187..1b72bda 100644
--- a/libgomp/testsuite/libgomp.oacc-c-c++-common/declare-3.c
+++ b/libgomp/testsuite/libgomp.oacc-c-c++-common/declare-3.c
@@ -6,7 +6,7 @@ 
 float *b;
 #pragma acc declare deviceptr (b)
 
-#pragma acc routine
+#pragma acc routine seq
 float *
 subr2 (void)
 {
diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/declare-4.c b/libgomp/testsuite/libgomp.oacc-c-c++-common/declare-4.c
index 36bf0eb..385b710 100644
--- a/libgomp/testsuite/libgomp.oacc-c-c++-common/declare-4.c
+++ b/libgomp/testsuite/libgomp.oacc-c-c++-common/declare-4.c
@@ -6,7 +6,7 @@ 
 float b;
 #pragma acc declare create (b)
 
-#pragma acc routine
+#pragma acc routine seq
 int
 func (int a)
 {
diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/host_data-1.c b/libgomp/testsuite/libgomp.oacc-c-c++-common/host_data-1.c
index fe843ec..9cb4665 100644
--- a/libgomp/testsuite/libgomp.oacc-c-c++-common/host_data-1.c
+++ b/libgomp/testsuite/libgomp.oacc-c-c++-common/host_data-1.c
@@ -8,7 +8,7 @@ 
 #include <cuda_runtime_api.h>
 #include <cublas_v2.h>
 
-#pragma acc routine
+#pragma acc routine seq
 void
 saxpy (int n, float a, float *x, float *y)
 {
diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/loop-default.h b/libgomp/testsuite/libgomp.oacc-c-c++-common/loop-default.h
index 55de04b..36e8497 100644
--- a/libgomp/testsuite/libgomp.oacc-c-c++-common/loop-default.h
+++ b/libgomp/testsuite/libgomp.oacc-c-c++-common/loop-default.h
@@ -3,7 +3,7 @@ 
 #include <string.h>
 #include <stdio.h>
 
-#pragma acc routine
+#pragma acc routine seq
 static int __attribute__ ((noinline)) coord ()
 {
   int res = 0;
diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/mode-transitions.c b/libgomp/testsuite/libgomp.oacc-c-c++-common/mode-transitions.c
index f62daf0..d3a5858 100644
--- a/libgomp/testsuite/libgomp.oacc-c-c++-common/mode-transitions.c
+++ b/libgomp/testsuite/libgomp.oacc-c-c++-common/mode-transitions.c
@@ -1122,7 +1122,7 @@  void t27()
 
 /* Check if worker-single variables get broadcastd to vectors.  */
 
-#pragma acc routine
+#pragma acc routine seq
 float t28_routine ()
 {
   return 2.71;
diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/parallel-loop-2.h b/libgomp/testsuite/libgomp.oacc-c-c++-common/parallel-loop-2.h
index 5691b7e..1934a2b 100644
--- a/libgomp/testsuite/libgomp.oacc-c-c++-common/parallel-loop-2.h
+++ b/libgomp/testsuite/libgomp.oacc-c-c++-common/parallel-loop-2.h
@@ -2,7 +2,7 @@ 
 #define VARS
 int a[1500];
 float b[10][15][10];
-#pragma acc routine
+#pragma acc routine seq
 __attribute__((noreturn)) void
 noreturn (void)
 {
diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/routine-1.c b/libgomp/testsuite/libgomp.oacc-c-c++-common/routine-1.c
index 0f70e26..e45e930 100644
--- a/libgomp/testsuite/libgomp.oacc-c-c++-common/routine-1.c
+++ b/libgomp/testsuite/libgomp.oacc-c-c++-common/routine-1.c
@@ -8,7 +8,7 @@ 
 
 #include <stdlib.h>
 
-#pragma acc routine
+#pragma acc routine seq
 TEMPLATE
 RETURN_1 fact(TYPE n) RETURN_2
 {
diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/routine-3.c b/libgomp/testsuite/libgomp.oacc-c-c++-common/routine-3.c
index 2cdd6bf..07b9551 100644
--- a/libgomp/testsuite/libgomp.oacc-c-c++-common/routine-3.c
+++ b/libgomp/testsuite/libgomp.oacc-c-c++-common/routine-3.c
@@ -6,7 +6,7 @@ 
 
 #include <stdlib.h>
 
-#pragma acc routine nohost
+#pragma acc routine nohost seq
 int
 foo (int n)
 {
diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/routine-bind-nohost-1.c b/libgomp/testsuite/libgomp.oacc-c-c++-common/routine-bind-nohost-1.c
index b991bb1..ec4a4a3 100644
--- a/libgomp/testsuite/libgomp.oacc-c-c++-common/routine-bind-nohost-1.c
+++ b/libgomp/testsuite/libgomp.oacc-c-c++-common/routine-bind-nohost-1.c
@@ -13,7 +13,7 @@ 
 /* "MINUS_TWO" is the device variant for function "TWO".  Similar for "THREE",
    and "FOUR".  Exercising different variants for declaring routines.  */
 
-#pragma acc routine nohost
+#pragma acc routine nohost seq
 extern int MINUS_TWO(void);
 
 int MINUS_TWO(void)
@@ -24,7 +24,7 @@  int MINUS_TWO(void)
 }
 
 extern int TWO(void);
-#pragma acc routine (TWO) bind(MINUS_TWO)
+#pragma acc routine (TWO) bind(MINUS_TWO) seq
 
 int TWO(void)
 {
@@ -34,7 +34,7 @@  int TWO(void)
 }
 
 
-#pragma acc routine nohost
+#pragma acc routine nohost seq
 int MINUS_THREE(void)
 {
   if (!acc_on_device(acc_device_not_host))
@@ -42,7 +42,7 @@  int MINUS_THREE(void)
   return -3;
 }
 
-#pragma acc routine bind(MINUS_THREE)
+#pragma acc routine bind(MINUS_THREE) seq
 extern int THREE(void);
 
 int THREE(void)
@@ -55,7 +55,7 @@  int THREE(void)
 
 /* Due to using a string in the bind clause, we don't need "MINUS_FOUR" in
    scope here.  */
-#pragma acc routine bind("MINUS_FOUR")
+#pragma acc routine bind("MINUS_FOUR") seq
 int FOUR(void)
 {
   if (acc_on_device(acc_device_not_host))
@@ -64,7 +64,7 @@  int FOUR(void)
 }
 
 extern int MINUS_FOUR(void);
-#pragma acc routine (MINUS_FOUR) nohost
+#pragma acc routine (MINUS_FOUR) nohost seq
 
 int MINUS_FOUR(void)
 {
diff --git a/libgomp/testsuite/libgomp.oacc-fortran/host_data-2.f90 b/libgomp/testsuite/libgomp.oacc-fortran/host_data-2.f90
index ff09218..4cdaa0c 100644
--- a/libgomp/testsuite/libgomp.oacc-fortran/host_data-2.f90
+++ b/libgomp/testsuite/libgomp.oacc-fortran/host_data-2.f90
@@ -81,7 +81,7 @@  subroutine saxpy (nn, aa, xx, yy)
   integer :: nn
   real*4 :: aa, xx(nn), yy(nn)
   integer i
-  !$acc routine
+  !$acc routine seq
 
   do i = 1, nn
     yy(i) = yy(i) + aa * xx(i)
diff --git a/libgomp/testsuite/libgomp.oacc-fortran/host_data-3.f b/libgomp/testsuite/libgomp.oacc-fortran/host_data-3.f
index 05ed949..fe0b904 100644
--- a/libgomp/testsuite/libgomp.oacc-fortran/host_data-3.f
+++ b/libgomp/testsuite/libgomp.oacc-fortran/host_data-3.f
@@ -67,7 +67,7 @@ 
       integer :: nn
       real*4 :: aa, xx(nn), yy(nn)
       integer i
-!$acc routine
+!$acc routine seq
 
       do i = 1, nn
          yy(i) = yy(i) + aa * xx(i)
diff --git a/libgomp/testsuite/libgomp.oacc-fortran/host_data-4.f90 b/libgomp/testsuite/libgomp.oacc-fortran/host_data-4.f90
index 6e379b5..e192c59 100644
--- a/libgomp/testsuite/libgomp.oacc-fortran/host_data-4.f90
+++ b/libgomp/testsuite/libgomp.oacc-fortran/host_data-4.f90
@@ -21,7 +21,7 @@  contains
     integer :: nn
     real*4 :: aa, xx(nn), yy(nn)
     integer i
-    !$acc routine
+    !$acc routine seq
 
     do i = 1, nn
        yy(i) = yy(i) + aa * xx(i)
diff --git a/libgomp/testsuite/libgomp.oacc-fortran/routine-1.f90 b/libgomp/testsuite/libgomp.oacc-fortran/routine-1.f90
index 3390515..3b85391 100644
--- a/libgomp/testsuite/libgomp.oacc-fortran/routine-1.f90
+++ b/libgomp/testsuite/libgomp.oacc-fortran/routine-1.f90
@@ -3,7 +3,7 @@ 
 
   interface
     recursive function fact (x)
-      !$acc routine
+      !$acc routine seq
       integer, intent(in) :: x
       integer :: fact
     end function fact
@@ -21,7 +21,7 @@ 
   end do
 end
 recursive function fact (x) result (res)
-  !$acc routine
+  !$acc routine seq
   integer, intent(in) :: x
   integer :: res
   if (x < 1) then
diff --git a/libgomp/testsuite/libgomp.oacc-fortran/routine-2.f90 b/libgomp/testsuite/libgomp.oacc-fortran/routine-2.f90
index 3d418b6..902ab69 100644
--- a/libgomp/testsuite/libgomp.oacc-fortran/routine-2.f90
+++ b/libgomp/testsuite/libgomp.oacc-fortran/routine-2.f90
@@ -4,7 +4,7 @@ 
   module m1
     contains
     recursive function fact (x) result (res)
-      !$acc routine
+      !$acc routine seq
       integer, intent(in) :: x
       integer :: res
       if (x < 1) then
diff --git a/libgomp/testsuite/libgomp.oacc-fortran/routine-3.f90 b/libgomp/testsuite/libgomp.oacc-fortran/routine-3.f90
index d233a63..c3f6edd 100644
--- a/libgomp/testsuite/libgomp.oacc-fortran/routine-3.f90
+++ b/libgomp/testsuite/libgomp.oacc-fortran/routine-3.f90
@@ -4,7 +4,7 @@ 
   integer, parameter :: n = 10
   integer :: a(n), i
   integer, external :: fact
-  !$acc routine (fact)
+  !$acc routine (fact) seq
   !$acc parallel
   !$acc loop
   do i = 1, n
@@ -16,7 +16,7 @@ 
   end do
 end
 recursive function fact (x) result (res)
-  !$acc routine
+  !$acc routine seq
   integer, intent(in) :: x
   integer :: res
   if (x < 1) then
diff --git a/libgomp/testsuite/libgomp.oacc-fortran/routine-4.f90 b/libgomp/testsuite/libgomp.oacc-fortran/routine-4.f90
index 3e5fb09..7e763b8 100644
--- a/libgomp/testsuite/libgomp.oacc-fortran/routine-4.f90
+++ b/libgomp/testsuite/libgomp.oacc-fortran/routine-4.f90
@@ -17,7 +17,7 @@ 
   end do
 end
 subroutine incr (x)
-  !$acc routine
+  !$acc routine seq
   integer, intent(inout) :: x
   x = x + 1
 end subroutine incr
diff --git a/libgomp/testsuite/libgomp.oacc-fortran/routine-5.f90 b/libgomp/testsuite/libgomp.oacc-fortran/routine-5.f90
index 956da8e..8c5d03a 100644
--- a/libgomp/testsuite/libgomp.oacc-fortran/routine-5.f90
+++ b/libgomp/testsuite/libgomp.oacc-fortran/routine-5.f90
@@ -15,7 +15,7 @@  program main
 contains
 
     function func (n) result (rc)
-    !$acc routine
+    !$acc routine seq
     integer, intent (in) :: n
     integer :: rc
 
diff --git a/libgomp/testsuite/libgomp.oacc-fortran/routine-9.f90 b/libgomp/testsuite/libgomp.oacc-fortran/routine-9.f90
index 95d1a13..6f05e28 100644
--- a/libgomp/testsuite/libgomp.oacc-fortran/routine-9.f90
+++ b/libgomp/testsuite/libgomp.oacc-fortran/routine-9.f90
@@ -6,7 +6,7 @@  program main
   integer, parameter :: n = 10
   integer :: a(n), i
   integer, external :: fact
-  !$acc routine (fact)
+  !$acc routine (fact) seq
   !$acc parallel
   !$acc loop
   do i = 1, n
@@ -20,7 +20,7 @@  end program main
 
 recursive function fact (x) result (res)
   implicit none
-  !$acc routine (fact)
+  !$acc routine (fact) seq
   integer, intent(in) :: x
   integer :: res
   if (x < 1) then