diff --git gcc/ChangeLog.gomp gcc/ChangeLog.gomp
index 06e8583..970e744 100644
--- gcc/ChangeLog.gomp
+++ gcc/ChangeLog.gomp
@@ -1,4 +1,11 @@
 2014-12-10  Thomas Schwinge  <thomas@codesourcery.com>
+
+	* omp-low.c (scan_omp_target): Remove taskreg_nesting_level and
+	target_nesting_level assertions.
+	(check_omp_nesting_restrictions): Rework OpenACC constructs
+	handling.  Update and extend the relevant test cases.
+
+2014-12-10  Thomas Schwinge  <thomas@codesourcery.com>
 	    Bernd Schmidt  <bernds@codesourcery.com>
 
 	* gimple.def (GIMPLE_OACC_KERNELS, GIMPLE_OACC_PARALLEL): Merge
diff --git gcc/omp-low.c gcc/omp-low.c
index 39e2f22..d16e2de 100644
--- gcc/omp-low.c
+++ gcc/omp-low.c
@@ -2647,12 +2647,6 @@ scan_omp_target (gimple stmt, omp_context *outer_ctx)
   tree name;
   bool offloaded = is_gimple_omp_offloaded (stmt);
 
-  if (is_gimple_omp_oacc_specifically (stmt))
-    {
-      gcc_assert (taskreg_nesting_level == 0);
-      gcc_assert (target_nesting_level == 0);
-    }
-
   ctx = new_omp_context (stmt, outer_ctx);
   ctx->field_map = splay_tree_new (splay_tree_compare_pointers, 0, 0);
   ctx->default_kind = OMP_CLAUSE_DEFAULT_SHARED;
@@ -2706,46 +2700,26 @@ scan_omp_teams (gimple stmt, omp_context *outer_ctx)
   scan_omp (gimple_omp_body_ptr (stmt), ctx);
 }
 
-/* Check OpenMP nesting restrictions.  */
+/* Check nesting restrictions.  */
 static bool
 check_omp_nesting_restrictions (gimple stmt, omp_context *ctx)
 {
-  /* TODO: While the OpenACC specification does allow for certain kinds of
-     nesting, we don't support many of these yet.  */
-  if (is_gimple_omp (stmt)
-      && is_gimple_omp_oacc_specifically (stmt))
+  /* TODO: Some OpenACC/OpenMP nesting should be allowed.  */
+
+  /* No nesting of non-OpenACC STMT (that is, an OpenMP one, or a GOMP builtin)
+     inside an OpenACC CTX.  */
+  if (!(is_gimple_omp (stmt)
+	&& is_gimple_omp_oacc_specifically (stmt)))
     {
-      /* Regular handling of OpenACC loop constructs.  */
-      if (gimple_code (stmt) == GIMPLE_OMP_FOR
-	  && gimple_omp_for_kind (stmt) == GF_OMP_FOR_KIND_OACC_LOOP)
-	goto cont;
-      /* No nesting of OpenACC STMT inside any OpenACC or OpenMP CTX different
-	 from an OpenACC data construct.  */
-      for (omp_context *ctx_ = ctx; ctx_ != NULL; ctx_ = ctx_->outer)
-	if (is_gimple_omp (ctx_->stmt)
-	    && !(gimple_code (ctx_->stmt) == GIMPLE_OMP_TARGET
-		 && (gimple_omp_target_kind (ctx_->stmt)
-		     == GF_OMP_TARGET_KIND_OACC_DATA)))
-	  {
-	    error_at (gimple_location (stmt),
-		      "may not be nested");
-	    return false;
-	  }
-    }
-  else
-    {
-      /* No nesting of non-OpenACC STMT (that is, an OpenMP one, or a GOMP
-	 builtin) inside any OpenACC CTX.  */
       for (omp_context *ctx_ = ctx; ctx_ != NULL; ctx_ = ctx_->outer)
 	if (is_gimple_omp (ctx_->stmt)
 	    && is_gimple_omp_oacc_specifically (ctx_->stmt))
 	  {
 	    error_at (gimple_location (stmt),
-		      "may not be nested");
+		      "non-OpenACC construct inside of OpenACC region");
 	    return false;
 	  }
     }
- cont:
 
   if (ctx != NULL)
     {
@@ -3003,20 +2977,74 @@ check_omp_nesting_restrictions (gimple stmt, omp_context *ctx)
       break;
     case GIMPLE_OMP_TARGET:
       for (; ctx != NULL; ctx = ctx->outer)
-	if (gimple_code (ctx->stmt) == GIMPLE_OMP_TARGET
-	    && gimple_omp_target_kind (ctx->stmt) == GF_OMP_TARGET_KIND_REGION)
-	  {
-	    const char *name;
-	    switch (gimple_omp_target_kind (stmt))
-	      {
-	      case GF_OMP_TARGET_KIND_REGION: name = "target"; break;
-	      case GF_OMP_TARGET_KIND_DATA: name = "target data"; break;
-	      case GF_OMP_TARGET_KIND_UPDATE: name = "target update"; break;
-	      default: gcc_unreachable ();
-	      }
-	    warning_at (gimple_location (stmt), 0,
-			"%s construct inside of target region", name);
-	  }
+	{
+	  if (gimple_code (ctx->stmt) != GIMPLE_OMP_TARGET)
+	    {
+	      if (is_gimple_omp (stmt)
+		  && is_gimple_omp_oacc_specifically (stmt)
+		  && is_gimple_omp (ctx->stmt))
+		{
+		  error_at (gimple_location (stmt),
+			    "OpenACC construct inside of non-OpenACC region");
+		  return false;
+		}
+	      continue;
+	    }
+
+	  const char *stmt_name, *ctx_stmt_name;
+	  switch (gimple_omp_target_kind (stmt))
+	    {
+	    case GF_OMP_TARGET_KIND_REGION: stmt_name = "target"; break;
+	    case GF_OMP_TARGET_KIND_DATA: stmt_name = "target data"; break;
+	    case GF_OMP_TARGET_KIND_UPDATE: stmt_name = "target update"; break;
+	    case GF_OMP_TARGET_KIND_OACC_PARALLEL: stmt_name = "parallel"; break;
+	    case GF_OMP_TARGET_KIND_OACC_KERNELS: stmt_name = "kernels"; break;
+	    case GF_OMP_TARGET_KIND_OACC_DATA: stmt_name = "data"; break;
+	    case GF_OMP_TARGET_KIND_OACC_UPDATE: stmt_name = "update"; break;
+	    case GF_OMP_TARGET_KIND_OACC_ENTER_EXIT_DATA: stmt_name = "enter/exit data"; break;
+	    default: gcc_unreachable ();
+	    }
+	  switch (gimple_omp_target_kind (ctx->stmt))
+	    {
+	    case GF_OMP_TARGET_KIND_REGION: ctx_stmt_name = "target"; break;
+	    case GF_OMP_TARGET_KIND_DATA: ctx_stmt_name = "target data"; break;
+	    case GF_OMP_TARGET_KIND_OACC_PARALLEL: ctx_stmt_name = "parallel"; break;
+	    case GF_OMP_TARGET_KIND_OACC_KERNELS: ctx_stmt_name = "kernels"; break;
+	    case GF_OMP_TARGET_KIND_OACC_DATA: ctx_stmt_name = "data"; break;
+	    default: gcc_unreachable ();
+	    }
+
+	  /* OpenACC/OpenMP mismatch?  */
+	  if (is_gimple_omp_oacc_specifically (stmt)
+	      != is_gimple_omp_oacc_specifically (ctx->stmt))
+	    {
+	      error_at (gimple_location (stmt),
+			"%s %s construct inside of %s %s region",
+			(is_gimple_omp_oacc_specifically (stmt)
+			 ? "OpenACC" : "OpenMP"), stmt_name,
+			(is_gimple_omp_oacc_specifically (ctx->stmt)
+			 ? "OpenACC" : "OpenMP"), ctx_stmt_name);
+	      return false;
+	    }
+	  if (is_gimple_omp_offloaded (ctx->stmt))
+	    {
+	      /* No GIMPLE_OMP_TARGET inside offloaded OpenACC CTX.  */
+	      if (is_gimple_omp_oacc_specifically (ctx->stmt))
+		{
+		  error_at (gimple_location (stmt),
+			    "%s construct inside of %s region",
+			    stmt_name, ctx_stmt_name);
+		  return false;
+		}
+	      else
+		{
+		  gcc_assert (!is_gimple_omp_oacc_specifically (stmt));
+		  warning_at (gimple_location (stmt), 0,
+			      "%s construct inside of %s region",
+			      stmt_name, ctx_stmt_name);
+		}
+	    }
+	}
       break;
     default:
       break;
diff --git gcc/testsuite/c-c++-common/goacc-gomp/nesting-fail-1.c gcc/testsuite/c-c++-common/goacc-gomp/nesting-fail-1.c
index 59f41a8..d52c7c0 100644
--- gcc/testsuite/c-c++-common/goacc-gomp/nesting-fail-1.c
+++ gcc/testsuite/c-c++-common/goacc-gomp/nesting-fail-1.c
@@ -1,5 +1,3 @@
-/* TODO: Some of these should either be allowed or fail with a more sensible
-   error message.  */
 void
 f_omp (void)
 {
@@ -7,24 +5,30 @@ f_omp (void)
 
 #pragma omp parallel
   {
-#pragma acc parallel	/* { dg-error "may not be nested" } */
+#pragma acc parallel /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
     ;
-#pragma acc kernels	/* { dg-error "may not be nested" } */
+#pragma acc kernels /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
     ;
-#pragma acc data	/* { dg-error "may not be nested" } */
+#pragma acc data /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
     ;
+#pragma acc update host(i) /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
+#pragma acc enter data copyin(i) /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
+#pragma acc exit data delete(i) /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
   }
 
 #pragma omp for
   for (i = 0; i < 3; i++)
     {
-#pragma acc parallel	/* { dg-error "may not be nested" } */
+#pragma acc parallel /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
       ;
-#pragma acc kernels	/* { dg-error "may not be nested" } */
+#pragma acc kernels /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
       ;
-#pragma acc data	/* { dg-error "may not be nested" } */
+#pragma acc data /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
       ;
-#pragma acc loop	/* { dg-error "may not be closely nested" } */
+#pragma acc update host(i) /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
+#pragma acc enter data copyin(i) /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
+#pragma acc exit data delete(i) /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
+#pragma acc loop /* { dg-error "may not be closely nested" } */
       for (i = 0; i < 2; ++i)
 	;
     }
@@ -32,22 +36,34 @@ f_omp (void)
 #pragma omp sections
   {
     {
-#pragma acc parallel	/* { dg-error "may not be nested" } */
+#pragma acc parallel /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
       ;
     }
 #pragma omp section
     {
-#pragma acc kernels	/* { dg-error "may not be nested" } */
+#pragma acc kernels /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
       ;
     }
 #pragma omp section
     {
-#pragma acc data	/* { dg-error "may not be nested" } */
+#pragma acc data /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
       ;
     }
 #pragma omp section
     {
-#pragma acc loop	/* { dg-error "may not be closely nested" } */
+#pragma acc update host(i) /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
+    }
+#pragma omp section
+    {
+#pragma acc enter data copyin(i) /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
+    }
+#pragma omp section
+    {
+#pragma acc exit data delete(i) /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
+    }
+#pragma omp section
+    {
+#pragma acc loop /* { dg-error "may not be closely nested" } */
       for (i = 0; i < 2; ++i)
 	;
     }
@@ -55,105 +71,121 @@ f_omp (void)
 
 #pragma omp single
   {
-#pragma acc parallel	/* { dg-error "may not be nested" } */
+#pragma acc parallel /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
     ;
-#pragma acc kernels	/* { dg-error "may not be nested" } */
+#pragma acc kernels /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
     ;
-#pragma acc data	/* { dg-error "may not be nested" } */
+#pragma acc data /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
     ;
-#pragma acc loop	/* { dg-error "may not be closely nested" } */
+#pragma acc update host(i) /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
+#pragma acc enter data copyin(i) /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
+#pragma acc exit data delete(i) /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
+#pragma acc loop /* { dg-error "may not be closely nested" } */
     for (i = 0; i < 2; ++i)
       ;
   }
 
 #pragma omp task
   {
-#pragma acc parallel	/* { dg-error "may not be nested" } */
+#pragma acc parallel /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
     ;
-#pragma acc kernels	/* { dg-error "may not be nested" } */
+#pragma acc kernels /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
     ;
-#pragma acc data	/* { dg-error "may not be nested" } */
+#pragma acc data /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
     ;
-#pragma acc loop	/* { dg-error "may not be closely nested" } */
+#pragma acc update host(i) /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
+#pragma acc enter data copyin(i) /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
+#pragma acc exit data delete(i) /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
+#pragma acc loop /* { dg-error "may not be closely nested" } */
     for (i = 0; i < 2; ++i)
       ;
   }
 
 #pragma omp master
   {
-#pragma acc parallel	/* { dg-error "may not be nested" } */
+#pragma acc parallel /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
     ;
-#pragma acc kernels	/* { dg-error "may not be nested" } */
+#pragma acc kernels /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
     ;
-#pragma acc data	/* { dg-error "may not be nested" } */
+#pragma acc data /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
     ;
-#pragma acc loop	/* { dg-error "may not be closely nested" } */
+#pragma acc update host(i) /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
+#pragma acc enter data copyin(i) /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
+#pragma acc exit data delete(i) /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
+#pragma acc loop /* { dg-error "may not be closely nested" } */
     for (i = 0; i < 2; ++i)
       ;
   }
 
 #pragma omp critical
   {
-#pragma acc parallel	/* { dg-error "may not be nested" } */
+#pragma acc parallel /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
     ;
-#pragma acc kernels	/* { dg-error "may not be nested" } */
+#pragma acc kernels /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
     ;
-#pragma acc data	/* { dg-error "may not be nested" } */
+#pragma acc data /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
     ;
-#pragma acc loop	/* { dg-error "may not be closely nested" } */
+#pragma acc update host(i) /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
+#pragma acc enter data copyin(i) /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
+#pragma acc exit data delete(i) /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
+#pragma acc loop /* { dg-error "may not be closely nested" } */
     for (i = 0; i < 2; ++i)
       ;
   }
 
 #pragma omp ordered
   {
-#pragma acc parallel	/* { dg-error "may not be nested" } */
+#pragma acc parallel /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
     ;
-#pragma acc kernels	/* { dg-error "may not be nested" } */
+#pragma acc kernels /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
     ;
-#pragma acc data	/* { dg-error "may not be nested" } */
+#pragma acc data /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
     ;
-#pragma acc loop	/* { dg-error "may not be closely nested" } */
+#pragma acc update host(i) /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
+#pragma acc enter data copyin(i) /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
+#pragma acc exit data delete(i) /* { dg-error "OpenACC construct inside of non-OpenACC region" } */
+#pragma acc loop /* { dg-error "may not be closely nested" } */
     for (i = 0; i < 2; ++i)
       ;
   }
 
 #pragma omp target
   {
-#pragma acc parallel	/* { dg-error "may not be nested" } */
+#pragma acc parallel /* { dg-error "OpenACC parallel construct inside of OpenMP target region" } */
     ;
-#pragma acc kernels	/* { dg-error "may not be nested" } */
+#pragma acc kernels /* { dg-error "OpenACC kernels construct inside of OpenMP target region" } */
     ;
-#pragma acc data	/* { dg-error "may not be nested" } */
+#pragma acc data /* { dg-error "OpenACC data construct inside of OpenMP target region" } */
     ;
+#pragma acc update host(i) /* { dg-error "OpenACC update construct inside of OpenMP target region" } */
+#pragma acc enter data copyin(i) /* { dg-error "OpenACC enter/exit data construct inside of OpenMP target region" } */
+#pragma acc exit data delete(i) /* { dg-error "OpenACC enter/exit data construct inside of OpenMP target region" } */
 #pragma acc loop
     for (i = 0; i < 2; ++i)
       ;
   }
 }
 
-/* TODO: Some of these should either be allowed or fail with a more sensible
-   error message.  */
 void
 f_acc_parallel (void)
 {
 #pragma acc parallel
   {
-#pragma omp parallel	/* { dg-error "may not be nested" } */
+#pragma omp parallel /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
     ;
   }
 
 #pragma acc parallel
   {
     int i;
-#pragma omp for		/* { dg-error "may not be nested" } */
+#pragma omp for /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
     for (i = 0; i < 3; i++)
       ;
   }
 
 #pragma acc parallel
   {
-#pragma omp sections	/* { dg-error "may not be nested" } */
+#pragma omp sections /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
     {
       ;
     }
@@ -161,25 +193,25 @@ f_acc_parallel (void)
 
 #pragma acc parallel
   {
-#pragma omp single	/* { dg-error "may not be nested" } */
+#pragma omp single /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
     ;
   }
 
 #pragma acc parallel
   {
-#pragma omp task	/* { dg-error "may not be nested" } */
+#pragma omp task /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
     ;
   }
 
 #pragma acc parallel
   {
-#pragma omp master	/* { dg-error "may not be nested" } */
+#pragma omp master /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
     ;
   }
 
 #pragma acc parallel
   {
-#pragma omp critical	/* { dg-error "may not be nested" } */
+#pragma omp critical /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
     ;
   }
 
@@ -187,44 +219,47 @@ f_acc_parallel (void)
   {
     int i;
 #pragma omp atomic write
-    i = 0;		/* { dg-error "may not be nested" } */
+    i = 0; /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
   }
 
 #pragma acc parallel
   {
-#pragma omp ordered	/* { dg-error "may not be nested" } */
+#pragma omp ordered /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
     ;
   }
 
 #pragma acc parallel
   {
-#pragma omp target	/* { dg-error "may not be nested" } */
+    int i;
+
+#pragma omp target /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
+    ;
+#pragma omp target data /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
     ;
+#pragma omp target update to(i) /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
   }
 }
 
-/* TODO: Some of these should either be allowed or fail with a more sensible
-   error message.  */
 void
 f_acc_kernels (void)
 {
 #pragma acc kernels
   {
-#pragma omp parallel	/* { dg-error "may not be nested" } */
+#pragma omp parallel /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
     ;
   }
 
 #pragma acc kernels
   {
     int i;
-#pragma omp for		/* { dg-error "may not be nested" } */
+#pragma omp for /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
     for (i = 0; i < 3; i++)
       ;
   }
 
 #pragma acc kernels
   {
-#pragma omp sections	/* { dg-error "may not be nested" } */
+#pragma omp sections /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
     {
       ;
     }
@@ -232,25 +267,25 @@ f_acc_kernels (void)
 
 #pragma acc kernels
   {
-#pragma omp single	/* { dg-error "may not be nested" } */
+#pragma omp single /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
     ;
   }
 
 #pragma acc kernels
   {
-#pragma omp task	/* { dg-error "may not be nested" } */
+#pragma omp task /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
     ;
   }
 
 #pragma acc kernels
   {
-#pragma omp master	/* { dg-error "may not be nested" } */
+#pragma omp master /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
     ;
   }
 
 #pragma acc kernels
   {
-#pragma omp critical	/* { dg-error "may not be nested" } */
+#pragma omp critical /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
     ;
   }
 
@@ -258,44 +293,47 @@ f_acc_kernels (void)
   {
     int i;
 #pragma omp atomic write
-    i = 0;		/* { dg-error "may not be nested" } */
+    i = 0; /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
   }
 
 #pragma acc kernels
   {
-#pragma omp ordered	/* { dg-error "may not be nested" } */
+#pragma omp ordered /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
     ;
   }
 
 #pragma acc kernels
   {
-#pragma omp target	/* { dg-error "may not be nested" } */
+    int i;
+
+#pragma omp target /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
+    ;
+#pragma omp target data /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
     ;
+#pragma omp target update to(i) /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
   }
 }
 
-/* TODO: Some of these should either be allowed or fail with a more sensible
-   error message.  */
 void
 f_acc_data (void)
 {
 #pragma acc data
   {
-#pragma omp parallel	/* { dg-error "may not be nested" } */
+#pragma omp parallel /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
     ;
   }
 
 #pragma acc data
   {
     int i;
-#pragma omp for		/* { dg-error "may not be nested" } */
+#pragma omp for /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
     for (i = 0; i < 3; i++)
       ;
   }
 
 #pragma acc data
   {
-#pragma omp sections	/* { dg-error "may not be nested" } */
+#pragma omp sections /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
     {
       ;
     }
@@ -303,25 +341,25 @@ f_acc_data (void)
 
 #pragma acc data
   {
-#pragma omp single	/* { dg-error "may not be nested" } */
+#pragma omp single /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
     ;
   }
 
 #pragma acc data
   {
-#pragma omp task	/* { dg-error "may not be nested" } */
+#pragma omp task /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
     ;
   }
 
 #pragma acc data
   {
-#pragma omp master	/* { dg-error "may not be nested" } */
+#pragma omp master /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
     ;
   }
 
 #pragma acc data
   {
-#pragma omp critical	/* { dg-error "may not be nested" } */
+#pragma omp critical /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
     ;
   }
 
@@ -329,24 +367,27 @@ f_acc_data (void)
   {
     int i;
 #pragma omp atomic write
-    i = 0;		/* { dg-error "may not be nested" } */
+    i = 0; /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
   }
 
 #pragma acc data
   {
-#pragma omp ordered	/* { dg-error "may not be nested" } */
+#pragma omp ordered /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
     ;
   }
 
 #pragma acc data
   {
-#pragma omp target	/* { dg-error "may not be nested" } */
+    int i;
+
+#pragma omp target /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
+    ;
+#pragma omp target data /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
     ;
+#pragma omp target update to(i) /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
   }
 }
 
-/* TODO: Some of these should either be allowed or fail with a more sensible
-   error message.  */
 void
 f_acc_loop (void)
 {
@@ -355,14 +396,14 @@ f_acc_loop (void)
 #pragma acc loop
   for (i = 0; i < 2; ++i)
     {
-#pragma omp parallel	/* { dg-error "may not be nested" } */
+#pragma omp parallel /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
       ;
     }
 
 #pragma acc loop
   for (i = 0; i < 2; ++i)
     {
-#pragma omp for		/* { dg-error "may not be nested" } */
+#pragma omp for /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
       for (i = 0; i < 3; i++)
 	;
     }
@@ -370,7 +411,7 @@ f_acc_loop (void)
 #pragma acc loop
   for (i = 0; i < 2; ++i)
     {
-#pragma omp sections	/* { dg-error "may not be nested" } */
+#pragma omp sections /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
       {
 	;
       }
@@ -379,28 +420,28 @@ f_acc_loop (void)
 #pragma acc loop
   for (i = 0; i < 2; ++i)
     {
-#pragma omp single	/* { dg-error "may not be nested" } */
+#pragma omp single /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
       ;
     }
 
 #pragma acc loop
   for (i = 0; i < 2; ++i)
     {
-#pragma omp task	/* { dg-error "may not be nested" } */
+#pragma omp task /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
       ;
     }
 
 #pragma acc loop
   for (i = 0; i < 2; ++i)
     {
-#pragma omp master	/* { dg-error "may not be nested" } */
+#pragma omp master /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
       ;
     }
 
 #pragma acc loop
   for (i = 0; i < 2; ++i)
     {
-#pragma omp critical	/* { dg-error "may not be nested" } */
+#pragma omp critical /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
       ;
     }
 
@@ -408,20 +449,23 @@ f_acc_loop (void)
   for (i = 0; i < 2; ++i)
     {
 #pragma omp atomic write
-      i = 0;		/* { dg-error "may not be nested" } */
+      i = 0; /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
     }
 
 #pragma acc loop
   for (i = 0; i < 2; ++i)
     {
-#pragma omp ordered	/* { dg-error "may not be nested" } */
+#pragma omp ordered /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
       ;
     }
 
 #pragma acc loop
   for (i = 0; i < 2; ++i)
     {
-#pragma omp target	/* { dg-error "may not be nested" } */
+#pragma omp target /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
       ;
+#pragma omp target data /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
+      ;
+#pragma omp target update to(i) /* { dg-error "non-OpenACC construct inside of OpenACC region" } */
     }
 }
diff --git gcc/testsuite/c-c++-common/goacc/nesting-1.c gcc/testsuite/c-c++-common/goacc/nesting-1.c
index a489d2d..4fbf018 100644
--- gcc/testsuite/c-c++-common/goacc/nesting-1.c
+++ gcc/testsuite/c-c++-common/goacc/nesting-1.c
@@ -46,11 +46,60 @@ f_acc_data (void)
 #pragma acc kernels
     ;
 
+#pragma acc kernels
+    {
+#pragma acc loop
+      for (i = 0; i < 2; ++i)
+	;
+    }
+
 #pragma acc data
     ;
 
+#pragma acc update host(i)
+
+#pragma acc enter data copyin(i)
+
+#pragma acc exit data delete(i)
+
 #pragma acc loop
     for (i = 0; i < 2; ++i)
       ;
+
+#pragma acc data
+    {
+#pragma acc parallel
+      ;
+
+#pragma acc parallel
+      {
+#pragma acc loop
+	for (i = 0; i < 2; ++i)
+	  ;
+      }
+
+#pragma acc kernels
+      ;
+
+#pragma acc kernels
+      {
+#pragma acc loop
+	for (i = 0; i < 2; ++i)
+	  ;
+      }
+
+#pragma acc data
+      ;
+
+#pragma acc update host(i)
+
+#pragma acc enter data copyin(i)
+
+#pragma acc exit data delete(i)
+
+#pragma acc loop
+      for (i = 0; i < 2; ++i)
+	;
+    }
   }
 }
diff --git gcc/testsuite/c-c++-common/goacc/nesting-2.c gcc/testsuite/c-c++-common/goacc/nesting-2.c
deleted file mode 100644
index 0d350c6..0000000
--- gcc/testsuite/c-c++-common/goacc/nesting-2.c
+++ /dev/null
@@ -1,11 +0,0 @@
-int i;
-
-void
-f_acc_data (void)
-{
-#pragma acc data
-  {
-#pragma acc update host(i)
-#pragma acc enter data copyin(i)
-  }
-}
diff --git gcc/testsuite/c-c++-common/goacc/nesting-fail-1.c gcc/testsuite/c-c++-common/goacc/nesting-fail-1.c
index 00dc602..a833806 100644
--- gcc/testsuite/c-c++-common/goacc/nesting-fail-1.c
+++ gcc/testsuite/c-c++-common/goacc/nesting-fail-1.c
@@ -5,12 +5,17 @@ f_acc_parallel (void)
 {
 #pragma acc parallel
   {
-#pragma acc parallel	/* { dg-error "may not be nested" } */
+    int i;
+
+#pragma acc parallel /* { dg-bogus "parallel construct inside of parallel region" "not implemented" { xfail *-*-* } } */
     ;
-#pragma acc kernels	/* { dg-error "may not be nested" } */
+#pragma acc kernels /* { dg-bogus "kernels construct inside of parallel region" "not implemented" { xfail *-*-* } } */
     ;
-#pragma acc data	/* { dg-error "may not be nested" } */
+#pragma acc data /* { dg-error "data construct inside of parallel region" } */
     ;
+#pragma acc update host(i) /* { dg-error "update construct inside of parallel region" } */
+#pragma acc enter data copyin(i) /* { dg-error "enter/exit data construct inside of parallel region" } */
+#pragma acc exit data delete(i) /* { dg-error "enter/exit data construct inside of parallel region" } */
   }
 }
 
@@ -21,11 +26,16 @@ f_acc_kernels (void)
 {
 #pragma acc kernels
   {
-#pragma acc parallel	/* { dg-error "may not be nested" } */
+    int i;
+
+#pragma acc parallel /* { dg-bogus "parallel construct inside of kernels region" "not implemented" { xfail *-*-* } } */
     ;
-#pragma acc kernels	/* { dg-error "may not be nested" } */
+#pragma acc kernels /* { dg-bogus "kernels construct inside of kernels region" "not implemented" { xfail *-*-* } } */
     ;
-#pragma acc data	/* { dg-error "may not be nested" } */
+#pragma acc data /* { dg-error "data construct inside of kernels region" } */
     ;
+#pragma acc update host(i) /* { dg-error "update construct inside of kernels region" } */
+#pragma acc enter data copyin(i) /* { dg-error "enter/exit data construct inside of kernels region" } */
+#pragma acc exit data delete(i) /* { dg-error "enter/exit data construct inside of kernels region" } */
   }
 }
diff --git gcc/testsuite/gfortran.dg/goacc/parallel-kernels-regions.f95 gcc/testsuite/gfortran.dg/goacc/parallel-kernels-regions.f95
index 33cb9cb..8b8e989 100644
--- gcc/testsuite/gfortran.dg/goacc/parallel-kernels-regions.f95
+++ gcc/testsuite/gfortran.dg/goacc/parallel-kernels-regions.f95
@@ -1,7 +1,7 @@
 ! { dg-do compile } 
 
-! OpenACC 2.0 allows nested parallel/kernels regions
-! However, in middle-end there is check for nested parallel
+! OpenACC 2.0 allows nested parallel/kernels regions, but this is not yet
+! supported.
 
 program test
   implicit none
@@ -9,48 +9,47 @@ program test
   integer :: i
 
   !$acc parallel
-    !$acc kernels 
+    !$acc kernels ! { dg-bogus "kernels construct inside of parallel region" "not implemented" { xfail *-*-* } }
     !$acc end kernels
   !$acc end parallel
 
   !$acc parallel
-    !$acc parallel ! { dg-error "may not be nested" }
+    !$acc parallel ! { dg-bogus "parallel construct inside of parallel region" "not implemented" { xfail *-*-* } }
     !$acc end parallel
   !$acc end parallel
 
   !$acc parallel
-    !$acc parallel ! { dg-error "may not be nested" }
+    !$acc parallel ! { dg-bogus "parallel construct inside of parallel region" "not implemented" { xfail *-*-* } }
     !$acc end parallel
-    !$acc kernels 
+    !$acc kernels ! { dg-bogus "kernels construct inside of parallel region" "not implemented" { xfail *-*-* } }
     !$acc end kernels
   !$acc end parallel
 
   !$acc kernels
-    !$acc kernels 
+    !$acc kernels ! { dg-bogus "kernels construct inside of kernels region" "not implemented" { xfail *-*-* } }
     !$acc end kernels
   !$acc end kernels
 
   !$acc kernels
-    !$acc parallel 
+    !$acc parallel ! { dg-bogus "parallel construct inside of kernels region" "not implemented" { xfail *-*-* } }
     !$acc end parallel
   !$acc end kernels
 
   !$acc kernels
-    !$acc parallel 
+    !$acc parallel ! { dg-bogus "parallel construct inside of kernels region" "not implemented" { xfail *-*-* } }
     !$acc end parallel
-    !$acc kernels 
+    !$acc kernels ! { dg-bogus "kernels construct inside of kernels region" "not implemented" { xfail *-*-* } }
     !$acc end kernels
   !$acc end kernels
 
   !$acc parallel
-    !$acc data ! { dg-error "may not be nested" }
+    !$acc data ! { dg-error "data construct inside of parallel region" }
     !$acc end data
   !$acc end parallel
 
   !$acc kernels
-    !$acc data
+    !$acc data ! { dg-error "data construct inside of kernels region" }
     !$acc end data
   !$acc end kernels
   
 end program test
-! { dg-prune-output "Error: may not be nested" }
