[OPENMP] Initial parsing/sema analysis of 'target' directive.

llvm-svn: 218110
diff --git a/clang/test/OpenMP/nesting_of_regions.cpp b/clang/test/OpenMP/nesting_of_regions.cpp
index eac2db2..203026b 100644
--- a/clang/test/OpenMP/nesting_of_regions.cpp
+++ b/clang/test/OpenMP/nesting_of_regions.cpp
@@ -86,6 +86,11 @@
 #pragma omp atomic
     ++a;
   }
+#pragma omp parallel
+  {
+#pragma omp target
+    ++a;
+  }
 
 // SIMD DIRECTIVE
 #pragma omp simd
@@ -197,6 +202,11 @@
 #pragma omp atomic // expected-error {{OpenMP constructs may not be nested inside a simd region}}
     ++a;
   }
+#pragma omp simd
+  for (int i = 0; i < 10; ++i) {
+#pragma omp target // expected-error {{OpenMP constructs may not be nested inside a simd region}}
+    ++a;
+  }
 
 // FOR DIRECTIVE
 #pragma omp for
@@ -331,6 +341,11 @@
 #pragma omp atomic
     ++a;
   }
+#pragma omp for
+  for (int i = 0; i < 10; ++i) {
+#pragma omp target
+    ++a;
+  }
 
 // FOR SIMD DIRECTIVE
 #pragma omp for simd
@@ -442,6 +457,11 @@
 #pragma omp atomic // expected-error {{OpenMP constructs may not be nested inside a simd region}}
     ++a;
   }
+#pragma omp for simd
+  for (int i = 0; i < 10; ++i) {
+#pragma omp target // expected-error {{OpenMP constructs may not be nested inside a simd region}}
+    ++a;
+  }
 
 // SECTIONS DIRECTIVE
 #pragma omp sections
@@ -583,6 +603,11 @@
 #pragma omp atomic
     ++a;
   }
+#pragma omp sections
+  {
+#pragma omp target
+    ++a;
+  }
 
 // SECTION DIRECTIVE
 #pragma omp section // expected-error {{orphaned 'omp section' directives are prohibited, it must be closely nested to a sections region}}
@@ -755,6 +780,12 @@
 #pragma omp atomic
     ++a;
   }
+#pragma omp sections
+  {
+#pragma omp section
+#pragma omp target
+    ++a;
+  }
 
 // SINGLE DIRECTIVE
 #pragma omp single
@@ -879,6 +910,11 @@
 #pragma omp atomic
     ++a;
   }
+#pragma omp single
+  {
+#pragma omp target
+    ++a;
+  }
 
 // MASTER DIRECTIVE
 #pragma omp master
@@ -1003,6 +1039,11 @@
 #pragma omp atomic
     ++a;
   }
+#pragma omp master
+  {
+#pragma omp target
+    ++a;
+  }
 
 // CRITICAL DIRECTIVE
 #pragma omp critical
@@ -1141,6 +1182,11 @@
 #pragma omp atomic
     ++a;
   }
+#pragma omp critical
+  {
+#pragma omp target
+    ++a;
+  }
 
 // PARALLEL FOR DIRECTIVE
 #pragma omp parallel for
@@ -1280,6 +1326,11 @@
 #pragma omp atomic
     ++a;
   }
+#pragma omp parallel for
+  for (int i = 0; i < 10; ++i) {
+#pragma omp target
+    ++a;
+  }
 
 // PARALLEL SECTIONS DIRECTIVE
 #pragma omp parallel sections
@@ -1410,6 +1461,11 @@
 #pragma omp atomic
     ++a;
   }
+#pragma omp parallel sections
+  {
+#pragma omp target
+    ++a;
+  }
 
 // TASK DIRECTIVE
 #pragma omp task
@@ -1488,6 +1544,11 @@
 #pragma omp atomic
     ++a;
   }
+#pragma omp task
+  {
+#pragma omp target
+    ++a;
+  }
 
 // ORDERED DIRECTIVE
 #pragma omp ordered
@@ -1602,38 +1663,43 @@
 #pragma omp atomic
     ++a;
   }
+#pragma omp ordered
+  {
+#pragma omp target
+    ++a;
+  }
 
 // ATOMIC DIRECTIVE
 #pragma omp atomic
-// expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
+  // expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
   {
 #pragma omp for // expected-error {{OpenMP constructs may not be nested inside an atomic region}}
     for (int i = 0; i < 10; ++i)
       ;
   }
 #pragma omp atomic
-// expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
+  // expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
   {
 #pragma omp simd // expected-error {{OpenMP constructs may not be nested inside an atomic region}}
     for (int i = 0; i < 10; ++i)
       ;
   }
 #pragma omp atomic
-// expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
+  // expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
   {
 #pragma omp for simd // expected-error {{OpenMP constructs may not be nested inside an atomic region}}
     for (int i = 0; i < 10; ++i)
       ;
   }
 #pragma omp atomic
-// expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
+  // expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
   {
 #pragma omp parallel // expected-error {{OpenMP constructs may not be nested inside an atomic region}}
     for (int i = 0; i < 10; ++i)
       ;
   }
 #pragma omp atomic
-// expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
+  // expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
   {
 #pragma omp sections // expected-error {{OpenMP constructs may not be nested inside an atomic region}}
     {
@@ -1641,7 +1707,7 @@
     }
   }
 #pragma omp atomic
-// expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
+  // expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
   {
 #pragma omp section // expected-error {{OpenMP constructs may not be nested inside an atomic region}}
     {
@@ -1649,7 +1715,7 @@
     }
   }
 #pragma omp atomic
-// expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
+  // expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
   {
 #pragma omp single // expected-error {{OpenMP constructs may not be nested inside an atomic region}}
     {
@@ -1657,7 +1723,7 @@
     }
   }
 #pragma omp atomic
-// expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
+  // expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
   {
 #pragma omp master // expected-error {{OpenMP constructs may not be nested inside an atomic region}}
     {
@@ -1665,7 +1731,7 @@
     }
   }
 #pragma omp atomic
-// expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
+  // expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
   {
 #pragma omp critical // expected-error {{OpenMP constructs may not be nested inside an atomic region}}
     {
@@ -1673,14 +1739,14 @@
     }
   }
 #pragma omp atomic
-// expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
+  // expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
   {
 #pragma omp parallel for // expected-error {{OpenMP constructs may not be nested inside an atomic region}}
     for (int i = 0; i < 10; ++i)
       ;
   }
 #pragma omp atomic
-// expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
+  // expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
   {
 #pragma omp parallel sections // expected-error {{OpenMP constructs may not be nested inside an atomic region}}
     {
@@ -1688,7 +1754,7 @@
     }
   }
 #pragma omp atomic
-// expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
+  // expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
   {
 #pragma omp task // expected-error {{OpenMP constructs may not be nested inside an atomic region}}
     {
@@ -1696,41 +1762,137 @@
     }
   }
 #pragma omp atomic
-// expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
+  // expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
   {
 #pragma omp taskyield // expected-error {{OpenMP constructs may not be nested inside an atomic region}}
     bar();
   }
 #pragma omp atomic
-// expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
+  // expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
   {
 #pragma omp barrier // expected-error {{OpenMP constructs may not be nested inside an atomic region}}
     bar();
   }
 #pragma omp atomic
-// expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
+  // expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
   {
 #pragma omp taskwait // expected-error {{OpenMP constructs may not be nested inside an atomic region}}
     bar();
   }
 #pragma omp atomic
-// expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
+  // expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
   {
 #pragma omp flush // expected-error {{OpenMP constructs may not be nested inside an atomic region}}
     bar();
   }
 #pragma omp atomic
-// expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
+  // expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
   {
 #pragma omp ordered // expected-error {{OpenMP constructs may not be nested inside an atomic region}}
     bar();
   }
 #pragma omp atomic
-// expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
+  // expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
   {
 #pragma omp atomic // expected-error {{OpenMP constructs may not be nested inside an atomic region}}
     ++a;
   }
+#pragma omp atomic
+  // expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
+  {
+#pragma omp target // expected-error {{OpenMP constructs may not be nested inside an atomic region}}
+    ++a;
+  }
+
+// TARGET DIRECTIVE
+#pragma omp target
+#pragma omp parallel
+  bar();
+#pragma omp target
+#pragma omp for
+  for (int i = 0; i < 10; ++i)
+    ;
+#pragma omp target
+#pragma omp simd
+  for (int i = 0; i < 10; ++i)
+    ;
+#pragma omp target
+#pragma omp for simd
+  for (int i = 0; i < 10; ++i)
+    ;
+#pragma omp target
+#pragma omp sections
+  {
+    bar();
+  }
+#pragma omp target
+#pragma omp section // expected-error {{'omp section' directive must be closely nested to a sections region, not a target region}}
+  {
+    bar();
+  }
+#pragma omp target
+#pragma omp single
+  bar();
+
+#pragma omp target
+#pragma omp master
+  {
+    bar();
+  }
+#pragma omp target
+#pragma omp critical
+  {
+    bar();
+  }
+#pragma omp target
+#pragma omp parallel for
+  for (int i = 0; i < 10; ++i)
+    ;
+#pragma omp target
+#pragma omp parallel sections
+  {
+    bar();
+  }
+#pragma omp target
+#pragma omp task
+  {
+    bar();
+  }
+#pragma omp target
+  {
+#pragma omp taskyield
+    bar();
+  }
+#pragma omp target
+  {
+#pragma omp barrier
+    bar();
+  }
+#pragma omp target
+  {
+#pragma omp taskwait
+    bar();
+  }
+#pragma omp target
+  {
+#pragma omp flush
+    bar();
+  }
+#pragma omp target
+  {
+#pragma omp ordered // expected-error {{region cannot be closely nested inside 'target' region; perhaps you forget to enclose 'omp ordered' directive into a for or a parallel for region with 'ordered' clause?}}
+    bar();
+  }
+#pragma omp target
+  {
+#pragma omp atomic
+    ++a;
+  }
+#pragma omp target
+  {
+#pragma omp target
+    ++a;
+  }
 }
 
 void foo() {
@@ -1816,6 +1978,11 @@
 #pragma omp atomic
     ++a;
   }
+#pragma omp parallel
+  {
+#pragma omp target
+    ++a;
+  }
 
 // SIMD DIRECTIVE
 #pragma omp simd
@@ -1920,6 +2087,11 @@
 #pragma omp atomic // expected-error {{OpenMP constructs may not be nested inside a simd region}}
     ++a;
   }
+#pragma omp simd
+  for (int i = 0; i < 10; ++i) {
+#pragma omp target // expected-error {{OpenMP constructs may not be nested inside a simd region}}
+    ++a;
+  }
 
 // FOR DIRECTIVE
 #pragma omp for
@@ -2044,6 +2216,11 @@
 #pragma omp atomic
     ++a;
   }
+#pragma omp for
+  for (int i = 0; i < 10; ++i) {
+#pragma omp target
+    ++a;
+  }
 
 // FOR SIMD DIRECTIVE
 #pragma omp for simd
@@ -2148,6 +2325,11 @@
 #pragma omp atomic // expected-error {{OpenMP constructs may not be nested inside a simd region}}
     ++a;
   }
+#pragma omp for simd
+  for (int i = 0; i < 10; ++i) {
+#pragma omp target // expected-error {{OpenMP constructs may not be nested inside a simd region}}
+    ++a;
+  }
 
 // SECTIONS DIRECTIVE
 #pragma omp sections
@@ -2264,6 +2446,11 @@
 #pragma omp atomic
     ++a;
   }
+#pragma omp sections
+  {
+#pragma omp target
+    ++a;
+  }
 
 // SECTION DIRECTIVE
 #pragma omp section // expected-error {{orphaned 'omp section' directives are prohibited, it must be closely nested to a sections region}}
@@ -2438,6 +2625,14 @@
       ++a;
     }
   }
+#pragma omp sections
+  {
+#pragma omp section
+    {
+#pragma omp target
+      ++a;
+    }
+  }
 
 // SINGLE DIRECTIVE
 #pragma omp single
@@ -2552,6 +2747,11 @@
 #pragma omp atomic
     ++a;
   }
+#pragma omp single
+  {
+#pragma omp target
+    ++a;
+  }
 
 // MASTER DIRECTIVE
 #pragma omp master
@@ -2676,6 +2876,11 @@
 #pragma omp atomic
     ++a;
   }
+#pragma omp master
+  {
+#pragma omp target
+    ++a;
+  }
 
 // CRITICAL DIRECTIVE
 #pragma omp critical
@@ -2819,6 +3024,11 @@
 #pragma omp atomic
     ++a;
   }
+#pragma omp critical
+  {
+#pragma omp target
+    ++a;
+  }
 
 // PARALLEL FOR DIRECTIVE
 #pragma omp parallel for
@@ -2958,6 +3168,11 @@
 #pragma omp atomic
     ++a;
   }
+#pragma omp parallel for
+  for (int i = 0; i < 10; ++i) {
+#pragma omp target
+    ++a;
+  }
 
 // PARALLEL SECTIONS DIRECTIVE
 #pragma omp parallel sections
@@ -3084,6 +3299,11 @@
 #pragma omp atomic
     ++a;
   }
+#pragma omp parallel sections
+  {
+#pragma omp target
+    ++a;
+  }
 
 // TASK DIRECTIVE
 #pragma omp task
@@ -3161,38 +3381,43 @@
 #pragma omp atomic
     ++a;
   }
+#pragma omp task
+  {
+#pragma omp target
+    ++a;
+  }
 
 // ATOMIC DIRECTIVE
 #pragma omp atomic
-// expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
+  // expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
   {
 #pragma omp for // expected-error {{OpenMP constructs may not be nested inside an atomic region}}
     for (int i = 0; i < 10; ++i)
       ;
   }
 #pragma omp atomic
-// expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
+  // expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
   {
 #pragma omp simd // expected-error {{OpenMP constructs may not be nested inside an atomic region}}
     for (int i = 0; i < 10; ++i)
       ;
   }
 #pragma omp atomic
-// expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
+  // expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
   {
 #pragma omp for simd // expected-error {{OpenMP constructs may not be nested inside an atomic region}}
     for (int i = 0; i < 10; ++i)
       ;
   }
 #pragma omp atomic
-// expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
+  // expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
   {
 #pragma omp parallel // expected-error {{OpenMP constructs may not be nested inside an atomic region}}
     for (int i = 0; i < 10; ++i)
       ;
   }
 #pragma omp atomic
-// expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
+  // expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
   {
 #pragma omp sections // expected-error {{OpenMP constructs may not be nested inside an atomic region}}
     {
@@ -3200,7 +3425,7 @@
     }
   }
 #pragma omp atomic
-// expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
+  // expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
   {
 #pragma omp section // expected-error {{OpenMP constructs may not be nested inside an atomic region}}
     {
@@ -3208,7 +3433,7 @@
     }
   }
 #pragma omp atomic
-// expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
+  // expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
   {
 #pragma omp single // expected-error {{OpenMP constructs may not be nested inside an atomic region}}
     {
@@ -3216,7 +3441,7 @@
     }
   }
 #pragma omp atomic
-// expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
+  // expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
   {
 #pragma omp master // expected-error {{OpenMP constructs may not be nested inside an atomic region}}
     {
@@ -3224,7 +3449,7 @@
     }
   }
 #pragma omp atomic
-// expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
+  // expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
   {
 #pragma omp critical // expected-error {{OpenMP constructs may not be nested inside an atomic region}}
     {
@@ -3232,14 +3457,14 @@
     }
   }
 #pragma omp atomic
-// expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
+  // expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
   {
 #pragma omp parallel for // expected-error {{OpenMP constructs may not be nested inside an atomic region}}
     for (int i = 0; i < 10; ++i)
       ;
   }
 #pragma omp atomic
-// expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
+  // expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
   {
 #pragma omp parallel sections // expected-error {{OpenMP constructs may not be nested inside an atomic region}}
     {
@@ -3247,7 +3472,7 @@
     }
   }
 #pragma omp atomic
-// expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
+  // expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
   {
 #pragma omp task // expected-error {{OpenMP constructs may not be nested inside an atomic region}}
     {
@@ -3255,41 +3480,137 @@
     }
   }
 #pragma omp atomic
-// expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
+  // expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
   {
 #pragma omp taskyield // expected-error {{OpenMP constructs may not be nested inside an atomic region}}
     bar();
   }
 #pragma omp atomic
-// expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
+  // expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
   {
 #pragma omp barrier // expected-error {{OpenMP constructs may not be nested inside an atomic region}}
     bar();
   }
 #pragma omp atomic
-// expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
+  // expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
   {
 #pragma omp taskwait // expected-error {{OpenMP constructs may not be nested inside an atomic region}}
     bar();
   }
 #pragma omp atomic
-// expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
+  // expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
   {
 #pragma omp flush // expected-error {{OpenMP constructs may not be nested inside an atomic region}}
     bar();
   }
 #pragma omp atomic
-// expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
+  // expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
   {
 #pragma omp ordered // expected-error {{OpenMP constructs may not be nested inside an atomic region}}
     bar();
   }
 #pragma omp atomic
-// expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
+  // expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
   {
 #pragma omp atomic // expected-error {{OpenMP constructs may not be nested inside an atomic region}}
     ++a;
   }
+#pragma omp atomic
+  // expected-error@+1 {{the statement for 'atomic' must be an expression statement of form '++x;', '--x;', 'x++;', 'x--;', 'x binop= expr;', 'x = x binop expr' or 'x = expr binop x', where x is an l-value expression with scalar type}}
+  {
+#pragma omp target // expected-error {{OpenMP constructs may not be nested inside an atomic region}}
+    ++a;
+  }
+
+// TARGET DIRECTIVE
+#pragma omp target
+#pragma omp parallel
+  bar();
+#pragma omp target
+#pragma omp for
+  for (int i = 0; i < 10; ++i)
+    ;
+#pragma omp target
+#pragma omp simd
+  for (int i = 0; i < 10; ++i)
+    ;
+#pragma omp target
+#pragma omp for simd
+  for (int i = 0; i < 10; ++i)
+    ;
+#pragma omp target
+#pragma omp sections
+  {
+    bar();
+  }
+#pragma omp target
+#pragma omp section // expected-error {{'omp section' directive must be closely nested to a sections region, not a target region}}
+  {
+    bar();
+  }
+#pragma omp target
+#pragma omp single
+  bar();
+
+#pragma omp target
+#pragma omp master
+  {
+    bar();
+  }
+#pragma omp target
+#pragma omp critical
+  {
+    bar();
+  }
+#pragma omp target
+#pragma omp parallel for
+  for (int i = 0; i < 10; ++i)
+    ;
+#pragma omp target
+#pragma omp parallel sections
+  {
+    bar();
+  }
+#pragma omp target
+#pragma omp task
+  {
+    bar();
+  }
+#pragma omp target
+  {
+#pragma omp taskyield
+    bar();
+  }
+#pragma omp target
+  {
+#pragma omp barrier
+    bar();
+  }
+#pragma omp target
+  {
+#pragma omp taskwait
+    bar();
+  }
+#pragma omp target
+  {
+#pragma omp flush
+    bar();
+  }
+#pragma omp target
+  {
+#pragma omp ordered // expected-error {{region cannot be closely nested inside 'target' region; perhaps you forget to enclose 'omp ordered' directive into a for or a parallel for region with 'ordered' clause?}}
+    bar();
+  }
+#pragma omp target
+  {
+#pragma omp atomic
+    ++a;
+  }
+#pragma omp target
+  {
+#pragma omp target
+    ++a;
+  }
   return foo<int>();
 }
 
diff --git a/clang/test/OpenMP/target_ast_print.cpp b/clang/test/OpenMP/target_ast_print.cpp
new file mode 100644
index 0000000..c57c049
--- /dev/null
+++ b/clang/test/OpenMP/target_ast_print.cpp
@@ -0,0 +1,57 @@
+// RUN: %clang_cc1 -verify -fopenmp=libiomp5 -ast-print %s | FileCheck %s
+// RUN: %clang_cc1 -fopenmp=libiomp5 -x c++ -std=c++11 -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp=libiomp5 -std=c++11 -include-pch %t -fsyntax-only -verify %s -ast-print | FileCheck %s
+// expected-no-diagnostics
+
+#ifndef HEADER
+#define HEADER
+
+void foo() {}
+
+template <typename T, int C>
+T tmain(T argc, T *argv) {
+#pragma omp target
+  foo();
+#pragma omp target if (argc > 0)
+  foo();
+#pragma omp target if (C)
+  foo();
+  return 0;
+}
+
+// CHECK: template <typename T = int, int C = 5> int tmain(int argc, int *argv) {
+// CHECK-NEXT: #pragma omp target
+// CHECK-NEXT: foo();
+// CHECK-NEXT: #pragma omp target if(argc > 0)
+// CHECK-NEXT: foo()
+// CHECK-NEXT: #pragma omp target if(5)
+// CHECK-NEXT: foo()
+// CHECK: template <typename T = char, int C = 1> char tmain(char argc, char *argv) {
+// CHECK-NEXT: #pragma omp target
+// CHECK-NEXT: foo();
+// CHECK-NEXT: #pragma omp target if(argc > 0)
+// CHECK-NEXT: foo()
+// CHECK-NEXT: #pragma omp target if(1)
+// CHECK-NEXT: foo()
+// CHECK: template <typename T, int C> T tmain(T argc, T *argv) {
+// CHECK-NEXT: #pragma omp target
+// CHECK-NEXT: foo();
+// CHECK-NEXT: #pragma omp target if(argc > 0)
+// CHECK-NEXT: foo()
+// CHECK-NEXT: #pragma omp target if(C)
+// CHECK-NEXT: foo()
+
+// CHECK-LABEL: int main(int argc, char **argv) {
+int main (int argc, char **argv) {
+#pragma omp target
+// CHECK-NEXT: #pragma omp target
+  foo();
+// CHECK-NEXT: foo();
+#pragma omp target if (argc > 0)
+// CHECK-NEXT: #pragma omp target if(argc > 0)
+  foo();
+// CHECK-NEXT: foo();
+  return tmain<int, 5>(argc, &argc) + tmain<char, 1>(argv[0][0], argv[0]);
+}
+
+#endif
diff --git a/clang/test/OpenMP/target_if_messages.cpp b/clang/test/OpenMP/target_if_messages.cpp
new file mode 100644
index 0000000..50a9ed2
--- /dev/null
+++ b/clang/test/OpenMP/target_if_messages.cpp
@@ -0,0 +1,46 @@
+// RUN: %clang_cc1 -verify -fopenmp=libiomp5 %s
+
+void foo() {
+}
+
+bool foobool(int argc) {
+  return argc;
+}
+
+struct S1; // expected-note {{declared here}}
+
+template <class T, class S> // expected-note {{declared here}}
+int tmain(T argc, S **argv) {
+  #pragma omp target if // expected-error {{expected '(' after 'if'}}
+  #pragma omp target if ( // expected-error {{expected expression}} expected-error {{expected ')'}} expected-note {{to match this '('}}
+  #pragma omp target if () // expected-error {{expected expression}}
+  #pragma omp target if (argc // expected-error {{expected ')'}} expected-note {{to match this '('}}
+  #pragma omp target if (argc)) // expected-warning {{extra tokens at the end of '#pragma omp target' are ignored}}
+  #pragma omp target if (argc > 0 ? argv[1] : argv[2])
+  #pragma omp target if (foobool(argc)), if (true) // expected-error {{directive '#pragma omp target' cannot contain more than one 'if' clause}}
+  #pragma omp target if (S) // expected-error {{'S' does not refer to a value}}
+  #pragma omp target if (argv[1]=2) // expected-error {{expected ')'}} expected-note {{to match this '('}}
+  #pragma omp target if (argc argc) // expected-error {{expected ')'}} expected-note {{to match this '('}}
+  #pragma omp target if(argc)
+  foo();
+
+  return 0;
+}
+
+int main(int argc, char **argv) {
+  #pragma omp target if // expected-error {{expected '(' after 'if'}}
+  #pragma omp target if ( // expected-error {{expected expression}} expected-error {{expected ')'}} expected-note {{to match this '('}}
+  #pragma omp target if () // expected-error {{expected expression}}
+  #pragma omp target if (argc // expected-error {{expected ')'}} expected-note {{to match this '('}}
+  #pragma omp target if (argc)) // expected-warning {{extra tokens at the end of '#pragma omp target' are ignored}}
+  #pragma omp target if (argc > 0 ? argv[1] : argv[2])
+  #pragma omp target if (foobool(argc)), if (true) // expected-error {{directive '#pragma omp target' cannot contain more than one 'if' clause}}
+  #pragma omp target if (S1) // expected-error {{'S1' does not refer to a value}}
+  #pragma omp target if (argv[1]=2) // expected-error {{expected ')'}} expected-note {{to match this '('}}
+  #pragma omp target if (argc argc) // expected-error {{expected ')'}} expected-note {{to match this '('}}
+  #pragma omp target if (1 0) // expected-error {{expected ')'}} expected-note {{to match this '('}}
+  #pragma omp target if(if(tmain(argc, argv) // expected-error {{expected expression}} expected-error {{expected ')'}} expected-note {{to match this '('}}
+  foo();
+
+  return tmain(argc, argv);
+}
diff --git a/clang/test/OpenMP/target_messages.cpp b/clang/test/OpenMP/target_messages.cpp
new file mode 100644
index 0000000..fb49228
--- /dev/null
+++ b/clang/test/OpenMP/target_messages.cpp
@@ -0,0 +1,64 @@
+// RUN: %clang_cc1 -verify -fopenmp=libiomp5 -std=c++11 -o - %s
+
+void foo() {
+}
+
+#pragma omp target // expected-error {{unexpected OpenMP directive '#pragma omp target'}}
+
+int main(int argc, char **argv) {
+  #pragma omp target { // expected-warning {{extra tokens at the end of '#pragma omp target' are ignored}}
+  foo();
+  #pragma omp target ( // expected-warning {{extra tokens at the end of '#pragma omp target' are ignored}}
+  foo();
+  #pragma omp target [ // expected-warning {{extra tokens at the end of '#pragma omp target' are ignored}}
+  foo();
+  #pragma omp target ] // expected-warning {{extra tokens at the end of '#pragma omp target' are ignored}}
+  foo();
+  #pragma omp target ) // expected-warning {{extra tokens at the end of '#pragma omp target' are ignored}}
+  foo();
+  #pragma omp target } // expected-warning {{extra tokens at the end of '#pragma omp target' are ignored}}
+  foo();
+  #pragma omp target
+  // expected-warning@+1 {{extra tokens at the end of '#pragma omp target' are ignored}}
+  #pragma omp target unknown()
+  foo();
+  L1:
+    foo();
+  #pragma omp target
+  ;
+  #pragma omp target
+  {
+    goto L1; // expected-error {{use of undeclared label 'L1'}}
+    argc++;
+  }
+
+  for (int i = 0; i < 10; ++i) {
+    switch(argc) {
+     case (0):
+      #pragma omp target
+      {
+        foo();
+        break; // expected-error {{'break' statement not in loop or switch statement}}
+        continue; // expected-error {{'continue' statement not in loop statement}}
+      }
+      default:
+       break;
+    }
+  }
+
+  goto L2; // expected-error {{use of undeclared label 'L2'}}
+  #pragma omp target
+  L2:
+  foo();
+  #pragma omp target
+  {
+    return 1; // expected-error {{cannot return from OpenMP region}}
+  }
+
+  [[]] // expected-error {{an attribute list cannot appear here}}
+  #pragma omp target
+  for (int n = 0; n < 100; ++n) {}
+
+  return 0;
+}
+