diff --git gcc/c-family/ChangeLog.gomp gcc/c-family/ChangeLog.gomp
index e092d53..3da377f 100644
--- gcc/c-family/ChangeLog.gomp
+++ gcc/c-family/ChangeLog.gomp
@@ -1,3 +1,8 @@
+2014-02-21  Thomas Schwinge  <thomas@codesourcery.com>
+
+	* c-pragma.c (oacc_pragmas): Add "data".
+	* c-pragma.h (enum pragma_kind): Add PRAGMA_OACC_DATA.
+
 2014-01-28  Thomas Schwinge  <thomas@codesourcery.com>
 
 	* c-pragma.h (pragma_omp_clause): Add PRAGMA_OMP_CLAUSE_COPY,
diff --git gcc/c-family/c-pragma.c gcc/c-family/c-pragma.c
index f69486a..08374aa 100644
--- gcc/c-family/c-pragma.c
+++ gcc/c-family/c-pragma.c
@@ -1169,6 +1169,7 @@ static vec<pragma_ns_name> registered_pp_pragmas;
 
 struct omp_pragma_def { const char *name; unsigned int id; };
 static const struct omp_pragma_def oacc_pragmas[] = {
+  { "data", PRAGMA_OACC_DATA },
   { "parallel", PRAGMA_OACC_PARALLEL },
 };
 static const struct omp_pragma_def omp_pragmas[] = {
diff --git gcc/c-family/c-pragma.h gcc/c-family/c-pragma.h
index 1ea5b1d..d092f9f 100644
--- gcc/c-family/c-pragma.h
+++ gcc/c-family/c-pragma.h
@@ -27,6 +27,7 @@ along with GCC; see the file COPYING3.  If not see
 typedef enum pragma_kind {
   PRAGMA_NONE = 0,
 
+  PRAGMA_OACC_DATA,
   PRAGMA_OACC_PARALLEL,
   PRAGMA_OMP_ATOMIC,
   PRAGMA_OMP_BARRIER,
diff --git gcc/c/ChangeLog.gomp gcc/c/ChangeLog.gomp
index b199957..9b95725 100644
--- gcc/c/ChangeLog.gomp
+++ gcc/c/ChangeLog.gomp
@@ -1,3 +1,11 @@
+2014-02-21  Thomas Schwinge  <thomas@codesourcery.com>
+
+	* c-parser.c (OACC_DATA_CLAUSE_MASK): New macro definition.
+	(c_parser_oacc_data): New function.
+	(c_parser_omp_construct): Handle PRAGMA_OACC_DATA.
+	* c-tree.h (c_finish_oacc_data): New prototype.
+	* c-typeck.c (c_finish_oacc_data): New function.
+
 2014-02-17  Thomas Schwinge  <thomas@codesourcery.com>
 
 	* c-parser.c (c_parser_omp_clause_name): Accept pcopy, pcopyin,
diff --git gcc/c/c-parser.c gcc/c/c-parser.c
index 7850eab..4643722 100644
--- gcc/c/c-parser.c
+++ gcc/c/c-parser.c
@@ -4776,10 +4776,14 @@ c_parser_label (c_parser *parser)
 
    openacc-construct:
      parallel-construct
+     data-construct
 
    parallel-construct:
      parallel-directive structured-block
 
+   data-construct:
+     data-directive structured-block
+
    OpenMP:
 
    statement:
@@ -11362,6 +11366,41 @@ c_parser_omp_structured_block (c_parser *parser)
 }
 
 /* OpenACC 2.0:
+   # pragma acc data oacc-data-clause[optseq] new-line
+     structured-block
+
+   LOC is the location of the #pragma token.
+*/
+
+#define OACC_DATA_CLAUSE_MASK					\
+	( (OMP_CLAUSE_MASK_1 << PRAGMA_OMP_CLAUSE_COPY)			\
+	| (OMP_CLAUSE_MASK_1 << PRAGMA_OMP_CLAUSE_COPYIN)		\
+	| (OMP_CLAUSE_MASK_1 << PRAGMA_OMP_CLAUSE_COPYOUT)		\
+	| (OMP_CLAUSE_MASK_1 << PRAGMA_OMP_CLAUSE_CREATE)		\
+	| (OMP_CLAUSE_MASK_1 << PRAGMA_OMP_CLAUSE_DEVICEPTR)		\
+	| (OMP_CLAUSE_MASK_1 << PRAGMA_OMP_CLAUSE_PRESENT)		\
+	| (OMP_CLAUSE_MASK_1 << PRAGMA_OMP_CLAUSE_PRESENT_OR_COPY)	\
+	| (OMP_CLAUSE_MASK_1 << PRAGMA_OMP_CLAUSE_PRESENT_OR_COPYIN)	\
+	| (OMP_CLAUSE_MASK_1 << PRAGMA_OMP_CLAUSE_PRESENT_OR_COPYOUT)	\
+	| (OMP_CLAUSE_MASK_1 << PRAGMA_OMP_CLAUSE_PRESENT_OR_CREATE) )
+
+static tree
+c_parser_oacc_data (location_t loc, c_parser *parser)
+{
+  tree stmt, clauses, block;
+
+  clauses =  c_parser_oacc_all_clauses (parser, OACC_DATA_CLAUSE_MASK,
+					"#pragma acc data");
+
+  block = c_begin_omp_parallel ();
+  add_stmt (c_parser_omp_structured_block (parser));
+
+  stmt = c_finish_oacc_data (loc, clauses, block);
+
+  return stmt;
+}
+
+/* OpenACC 2.0:
    # pragma acc parallel oacc-parallel-clause[optseq] new-line
      structured-block
 
@@ -13675,6 +13714,9 @@ c_parser_omp_construct (c_parser *parser)
 
   switch (p_kind)
     {
+    case PRAGMA_OACC_DATA:
+      stmt = c_parser_oacc_data (loc, parser);
+      break;
     case PRAGMA_OACC_PARALLEL:
       stmt = c_parser_oacc_parallel (loc, parser);
       break;
diff --git gcc/c/c-tree.h gcc/c/c-tree.h
index c174c7a..c84d3d7 100644
--- gcc/c/c-tree.h
+++ gcc/c/c-tree.h
@@ -634,6 +634,7 @@ extern tree c_finish_goto_label (location_t, tree);
 extern tree c_finish_goto_ptr (location_t, tree);
 extern tree c_expr_to_decl (tree, bool *, bool *);
 extern tree c_finish_oacc_parallel (location_t, tree, tree);
+extern tree c_finish_oacc_data (location_t, tree, tree);
 extern tree c_begin_omp_parallel (void);
 extern tree c_finish_omp_parallel (location_t, tree, tree);
 extern tree c_begin_omp_task (void);
diff --git gcc/c/c-typeck.c gcc/c/c-typeck.c
index 76d655b..8c4445b 100644
--- gcc/c/c-typeck.c
+++ gcc/c/c-typeck.c
@@ -11122,6 +11122,25 @@ c_finish_oacc_parallel (location_t loc, tree clauses, tree block)
   return add_stmt (stmt);
 }
 
+/* Generate OACC_DATA, with CLAUSES and BLOCK as its compound
+   statement.  LOC is the location of the OACC_DATA.  */
+
+tree
+c_finish_oacc_data (location_t loc, tree clauses, tree block)
+{
+  tree stmt;
+
+  block = c_end_compound_stmt (loc, block, true);
+
+  stmt = make_node (OACC_DATA);
+  TREE_TYPE (stmt) = void_type_node;
+  OACC_DATA_CLAUSES (stmt) = clauses;
+  OACC_DATA_BODY (stmt) = block;
+  SET_EXPR_LOCATION (stmt, loc);
+
+  return add_stmt (stmt);
+}
+
 /* Like c_begin_compound_stmt, except force the retention of the BLOCK.  */
 
 tree
diff --git gcc/testsuite/ChangeLog.gomp gcc/testsuite/ChangeLog.gomp
index fbccfa3..41d73b6 100644
--- gcc/testsuite/ChangeLog.gomp
+++ gcc/testsuite/ChangeLog.gomp
@@ -1,3 +1,13 @@
+2014-02-21  Thomas Schwinge  <thomas@codesourcery.com>
+
+	* c-c++-common/goacc-gomp/nesting-fail-1.c: Extend for OpenACC
+	data construct.
+	* c-c++-common/goacc/nesting-fail-1.c: Likewise.
+	* c-c++-common/goacc/parallel-fail-1.c: Rename to...
+	* c-c++-common/goacc/clauses-fail.c: ... this new file.  Extend
+	for OpenACC data construct.
+	* c-c++-common/goacc/data-1.c: New file.
+
 2014-02-18  Thomas Schwinge  <thomas@codesourcery.com>
 
 	* gcc.dg/goacc/parallel-sb-1.c: New file.
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 875ec66..78fb45b 100644
--- gcc/testsuite/c-c++-common/goacc-gomp/nesting-fail-1.c
+++ gcc/testsuite/c-c++-common/goacc-gomp/nesting-fail-1.c
@@ -1,7 +1,7 @@
 /* TODO: Some of these should either be allowed or fail with a more sensible
    error message.  */
 void
-f1 (void)
+f_omp (void)
 {
   int i;
 
@@ -9,6 +9,8 @@ f1 (void)
   {
 #pragma acc parallel	/* { dg-error "may not be nested" } */
     ;
+#pragma acc data	/* { dg-error "may not be nested" } */
+    ;
   }
 
 #pragma omp for
@@ -16,49 +18,68 @@ f1 (void)
     {
 #pragma acc parallel	/* { dg-error "may not be nested" } */
       ;
+#pragma acc data	/* { dg-error "may not be nested" } */
+      ;
     }
 
 #pragma omp sections
   {
+    {
 #pragma acc parallel	/* { dg-error "may not be nested" } */
-    ;
+      ;
+    }
+#pragma omp section
+    {
+#pragma acc data	/* { dg-error "may not be nested" } */
+      ;
+    }
   }
 
 #pragma omp single
   {
 #pragma acc parallel	/* { dg-error "may not be nested" } */
     ;
+#pragma acc data	/* { dg-error "may not be nested" } */
+    ;
   }
 
 #pragma omp task
   {
 #pragma acc parallel	/* { dg-error "may not be nested" } */
     ;
+#pragma acc data	/* { dg-error "may not be nested" } */
+    ;
   }
 
 #pragma omp master
   {
 #pragma acc parallel	/* { dg-error "may not be nested" } */
     ;
+#pragma acc data	/* { dg-error "may not be nested" } */
+    ;
   }
 
 #pragma omp critical
   {
 #pragma acc parallel	/* { dg-error "may not be nested" } */
     ;
+#pragma acc data	/* { dg-error "may not be nested" } */
+    ;
   }
 
 #pragma omp ordered
   {
 #pragma acc parallel	/* { dg-error "may not be nested" } */
     ;
+#pragma acc data	/* { dg-error "may not be nested" } */
+    ;
   }
 }
 
 /* TODO: Some of these should either be allowed or fail with a more sensible
    error message.  */
 void
-f2 (void)
+f_acc_parallel (void)
 {
 #pragma acc parallel
   {
@@ -119,3 +140,68 @@ f2 (void)
     ;
   }
 }
+
+/* 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 acc data
+  {
+    int i;
+#pragma omp for		/* { dg-error "may not be nested" } */
+    for (i = 0; i < 3; i++)
+      ;
+  }
+
+#pragma acc data
+  {
+#pragma omp sections	/* { dg-error "may not be nested" } */
+    {
+      ;
+    }
+  }
+
+#pragma acc data
+  {
+#pragma omp single	/* { dg-error "may not be nested" } */
+    ;
+  }
+
+#pragma acc data
+  {
+#pragma omp task	/* { dg-error "may not be nested" } */
+    ;
+  }
+
+#pragma acc data
+  {
+#pragma omp master	/* { dg-error "may not be nested" } */
+    ;
+  }
+
+#pragma acc data
+  {
+#pragma omp critical	/* { dg-error "may not be nested" } */
+    ;
+  }
+
+#pragma acc data
+  {
+    int i;
+#pragma omp atomic write
+    i = 0;		/* { dg-error "may not be nested" } */
+  }
+
+#pragma acc data
+  {
+#pragma omp ordered	/* { dg-error "may not be nested" } */
+    ;
+  }
+}
diff --git gcc/testsuite/c-c++-common/goacc/clauses-fail.c gcc/testsuite/c-c++-common/goacc/clauses-fail.c
new file mode 100644
index 0000000..b0dd042
--- /dev/null
+++ gcc/testsuite/c-c++-common/goacc/clauses-fail.c
@@ -0,0 +1,9 @@
+void
+f (void)
+{
+#pragma acc parallel one /* { dg-error "expected clause before 'one'" } */
+  ;
+
+#pragma acc data two /* { dg-error "expected clause before 'two'" } */
+  ;
+}
diff --git gcc/testsuite/c-c++-common/goacc/data-1.c gcc/testsuite/c-c++-common/goacc/data-1.c
new file mode 100644
index 0000000..8094575
--- /dev/null
+++ gcc/testsuite/c-c++-common/goacc/data-1.c
@@ -0,0 +1,6 @@
+void
+foo (void)
+{
+#pragma acc data
+  ;
+}
diff --git gcc/testsuite/c-c++-common/goacc/nesting-fail-1.c gcc/testsuite/c-c++-common/goacc/nesting-fail-1.c
index 6501397..24a4c11 100644
--- gcc/testsuite/c-c++-common/goacc/nesting-fail-1.c
+++ gcc/testsuite/c-c++-common/goacc/nesting-fail-1.c
@@ -1,11 +1,27 @@
 /* TODO: While the OpenACC specification does allow for certain kinds of
    nesting, we don't support that yet.  */
 void
-f1 (void)
+f_acc_parallel (void)
 {
 #pragma acc parallel
   {
 #pragma acc parallel	/* { dg-error "may not be nested" } */
     ;
+#pragma acc data	/* { dg-error "may not be nested" } */
+    ;
+  }
+}
+
+/* TODO: While the OpenACC specification does allow for certain kinds of
+   nesting, we don't support that yet.  */
+void
+f_acc_data (void)
+{
+#pragma acc data
+  {
+#pragma acc parallel	/* { dg-error "may not be nested" } */
+    ;
+#pragma acc data	/* { dg-error "may not be nested" } */
+    ;
   }
 }
diff --git gcc/testsuite/c-c++-common/goacc/parallel-fail-1.c gcc/testsuite/c-c++-common/goacc/parallel-fail-1.c
deleted file mode 100644
index efc6f14..0000000
--- gcc/testsuite/c-c++-common/goacc/parallel-fail-1.c
+++ /dev/null
@@ -1,6 +0,0 @@
-void
-foo (void)
-{
-#pragma acc parallel foo	/* { dg-error "expected clause before 'foo'" } */
-  foo ();
-}
diff --git libgomp/ChangeLog.gomp libgomp/ChangeLog.gomp
index 5c15656..b90b09b 100644
--- libgomp/ChangeLog.gomp
+++ libgomp/ChangeLog.gomp
@@ -1,5 +1,7 @@
 2014-02-21  Thomas Schwinge  <thomas@codesourcery.com>
 
+	* testsuite/libgomp.oacc-c/data-1.c: New file.
+
 	* libgomp.map (GOACC_2.0): Add GOACC_data_end, GOACC_data_start.
 	* libgomp_g.h (GOACC_data_start, GOACC_data_end): New prototypes.
 	* oacc-parallel.c (GOACC_data_start, GOACC_data_end): New
diff --git libgomp/testsuite/libgomp.oacc-c/data-1.c libgomp/testsuite/libgomp.oacc-c/data-1.c
new file mode 100644
index 0000000..8f9a17a
--- /dev/null
+++ libgomp/testsuite/libgomp.oacc-c/data-1.c
@@ -0,0 +1,170 @@
+/* { dg-do run } */
+
+extern void abort ();
+
+int i;
+
+int main(void)
+{
+  int j;
+
+#if 0
+  i = -1;
+  j = -2;
+#pragma acc data copyin (i, j)
+  {
+    // TODO: check that variables have been mapped.
+    if (i != -1 || j != -2)
+      abort ();
+    i = 2;
+    j = 1;
+    if (i != 2 || j != 1)
+      abort ();
+  }
+  if (i != 2 || j != 1)
+    abort ();
+
+  i = -1;
+  j = -2;
+#pragma acc data copyout (i, j)
+  {
+    // TODO: check that variables have been mapped.
+    if (i != -1 || j != -2)
+      abort ();
+    i = 2;
+    j = 1;
+    if (i != 2 || j != 1)
+      abort ();
+  }
+  if (i != -1 || j != -2)
+    abort ();
+
+  i = -1;
+  j = -2;
+#pragma acc data copy (i, j)
+  {
+    // TODO: check that variables have been mapped.
+    if (i != -1 || j != -2)
+      abort ();
+    i = 2;
+    j = 1;
+    if (i != 2 || j != 1)
+      abort ();
+  }
+  if (i != -1 || j != -2)
+    abort ();
+
+  i = -1;
+  j = -2;
+#pragma acc data create (i, j)
+  {
+    // TODO: check that variables have been mapped.
+    if (i != -1 || j != -2)
+      abort ();
+    i = 2;
+    j = 1;
+    if (i != 2 || j != 1)
+      abort ();
+  }
+  if (i != -1 || j != -2)
+    abort ();
+#endif
+
+  i = -1;
+  j = -2;
+#pragma acc data present_or_copyin (i, j)
+  {
+    // TODO: check that variables have been mapped.
+    if (i != -1 || j != -2)
+      abort ();
+    i = 2;
+    j = 1;
+    if (i != 2 || j != 1)
+      abort ();
+  }
+  if (i != 2 || j != 1)
+    abort ();
+
+#if 0
+  i = -1;
+  j = -2;
+#pragma acc data present_or_copyout (i, j)
+  {
+    // TODO: check that variables have been mapped.
+    if (i != -1 || j != -2)
+      abort ();
+    i = 2;
+    j = 1;
+    if (i != 2 || j != 1)
+      abort ();
+  }
+  if (i != -1 || j != -2)
+    abort ();
+#endif
+
+  i = -1;
+  j = -2;
+#pragma acc data present_or_copy (i, j)
+  {
+    // TODO: check that variables have been mapped.
+    if (i != -1 || j != -2)
+      abort ();
+    i = 2;
+    j = 1;
+    if (i != 2 || j != 1)
+      abort ();
+  }
+  if (i != -1 || j != -2)
+    abort ();
+
+#if 0
+  i = -1;
+  j = -2;
+#pragma acc data present_or_create (i, j)
+  {
+    // TODO: check that variables have been mapped.
+    i = 2;
+    j = 1;
+    if (i != 2 || j != 1)
+      abort ();
+  }
+  if (i != -1 || j != -2)
+    abort ();
+#endif
+
+#if 0
+  i = -1;
+  j = -2;
+#pragma acc data present (i, j)
+  {
+    // TODO: check that variables have been mapped.
+    if (i != -1 || j != -2)
+      abort ();
+    i = 2;
+    j = 1;
+    if (i != 2 || j != 1)
+      abort ();
+  }
+  if (i != -1 || j != -2)
+    abort ();
+#endif
+
+#if 0
+  i = -1;
+  j = -2;
+#pragma acc data
+  {
+    // TODO: check that variables have been mapped.
+    if (i != -1 || j != -2)
+      abort ();
+    i = 2;
+    j = 1;
+    if (i != 2 || j != 1)
+      abort ();
+  }
+  if (i != -1 || j != -2)
+    abort ();
+#endif
+
+  return 0;
+}
