[PR c/64765, c/64880] Support OpenACC Combined Directives in C, C++

	gcc/c-family/
	PR c/64765
	PR c/64880
	* c-common.h (c_oacc_split_loop_clauses): Declare function.
	* c-omp.c (c_oacc_split_loop_clauses): New function.
	gcc/c/
	PR c/64765
	PR c/64880
	* c-parser.c (c_parser_oacc_loop): Add mask, cclauses formal
	parameters, and handle these.  Adjust all users.
	(c_parser_oacc_kernels, c_parser_oacc_parallel): Merge functions
	into...
	(c_parser_oacc_kernels_parallel): ... this new function.  Adjust
	all users.
	* c-tree.h (c_finish_oacc_parallel, c_finish_oacc_kernels): Don't
	declare functions.
	(c_finish_omp_construct): Declare function.
	* c-typeck.c (c_finish_oacc_parallel, c_finish_oacc_kernels):
	Merge functions into...
	(c_finish_omp_construct): ... this new function.
	gcc/cp/
	PR c/64765
	PR c/64880
	* cp-tree.h (finish_oacc_kernels, finish_oacc_parallel): Don't
	declare functions.
	(finish_omp_construct): Declare function.
	* parser.c (cp_parser_oacc_loop): Add p_name, mask, cclauses
	formal parameters, and handle these.  Adjust all users.
	(cp_parser_oacc_kernels, cp_parser_oacc_parallel): Merge functions
	into...
	(cp_parser_oacc_kernels_parallel): ... this new function.  Adjust
	all users.
	* semantics.c (finish_oacc_kernels, finish_oacc_parallel): Merge functions into...
	(finish_omp_construct): ... this new function.
	gcc/
	* tree.h (OACC_PARALLEL_BODY, OACC_PARALLEL_CLAUSES)
	(OACC_KERNELS_BODY, OACC_KERNELS_CLAUSES, OACC_KERNELS_COMBINED)
	(OACC_PARALLEL_COMBINED): Don't define macros.  Adjust all users.
	gcc/testsuite/
	PR c/64765
	PR c/64880
	* c-c++-common/goacc/loop-1.c: Don't skip for C++.  Don't prune
	sorry message.
	(PR64765): New function.
	* gfortran.dg/goacc/coarray_2.f90: XFAIL.
	* gfortran.dg/goacc/combined_loop.f90: Extend.  Don't prune
	sorry message.
	* gfortran.dg/goacc/cray.f95: Refine prune directive.
	* gfortran.dg/goacc/parameter.f95: Likewise.
	libgomp/
	* testsuite/libgomp.oacc-c-c++-common/combdir-1.c: New file.
	* testsuite/libgomp.oacc-fortran/combdir-1.f90: Likewise.

From-SVN: r229404

4 files changed, 122 insertions, 100 deletions

diff --git a/gcc/cp/ChangeLog b/gcc/cp/ChangeLog
index e900e29..979ebb9 100644
--- a/gcc/cp/ChangeLog
+++ b/gcc/cp/ChangeLog
@@ -1,3 +1,21 @@
+2015-10-27  Thomas Schwinge  <thomas@codesourcery.com>
+	    James Norris  <jnorris@codesourcery.com>
+	    Cesar Philippidis  <cesar@codesourcery.com>
+
+	PR c/64765
+	PR c/64880
+	* cp-tree.h (finish_oacc_kernels, finish_oacc_parallel): Don't
+	declare functions.
+	(finish_omp_construct): Declare function.
+	* parser.c (cp_parser_oacc_loop): Add p_name, mask, cclauses
+	formal parameters, and handle these.  Adjust all users.
+	(cp_parser_oacc_kernels, cp_parser_oacc_parallel): Merge functions
+	into...
+	(cp_parser_oacc_kernels_parallel): ... this new function.  Adjust
+	all users.
+	* semantics.c (finish_oacc_kernels, finish_oacc_parallel): Merge functions into...
+	(finish_omp_construct): ... this new function.
+
 2015-10-25  Jason Merrill  <jason@redhat.com>
 
 	DR 2179
diff --git a/gcc/cp/cp-tree.h b/gcc/cp/cp-tree.h
index 16db41f..af2ba64 100644
--- a/gcc/cp/cp-tree.h
+++ b/gcc/cp/cp-tree.h
@@ -6318,8 +6318,7 @@ extern void finish_omp_threadprivate		(tree);
 extern tree begin_omp_structured_block		(void);
 extern tree finish_omp_structured_block		(tree);
 extern tree finish_oacc_data			(tree, tree);
-extern tree finish_oacc_kernels			(tree, tree);
-extern tree finish_oacc_parallel		(tree, tree);
+extern tree finish_omp_construct		(enum tree_code, tree, tree);
 extern tree begin_omp_parallel			(void);
 extern tree finish_omp_parallel			(tree, tree);
 extern tree begin_omp_task			(void);
diff --git a/gcc/cp/parser.c b/gcc/cp/parser.c
index 7555bf3..0354029 100644
--- a/gcc/cp/parser.c
+++ b/gcc/cp/parser.c
@@ -34298,69 +34298,64 @@ cp_parser_oacc_enter_exit_data (cp_parser *parser, cp_token *pragma_tok,
 }
 
 /* OpenACC 2.0:
-   # pragma acc kernels oacc-kernels-clause[optseq] new-line
-     structured-block  */
-
-#define OACC_KERNELS_CLAUSE_MASK					\
-	( (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_ASYNC)		\
-	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_COPY)		\
-	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_COPYIN)		\
-	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_COPYOUT)		\
-	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_CREATE)		\
-	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_DEVICEPTR)		\
-	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_IF)			\
-	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_PRESENT)		\
-	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_PRESENT_OR_COPY)	\
-	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_PRESENT_OR_COPYIN)	\
-	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_PRESENT_OR_COPYOUT)	\
-	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_PRESENT_OR_CREATE)	\
-	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_WAIT))
-
-static tree
-cp_parser_oacc_kernels (cp_parser *parser, cp_token *pragma_tok)
-{
-  tree stmt, clauses, block;
-  unsigned int save;
-
-  clauses = cp_parser_oacc_all_clauses (parser, OACC_KERNELS_CLAUSE_MASK,
-					"#pragma acc kernels", pragma_tok);
-
-  block = begin_omp_parallel ();
-  save = cp_parser_begin_omp_structured_block (parser);
-  cp_parser_statement (parser, NULL_TREE, false, NULL);
-  cp_parser_end_omp_structured_block (parser, save);
-  stmt = finish_oacc_kernels (clauses, block);
-  return stmt;
-}
-
-/* OpenACC 2.0:
    # pragma acc loop oacc-loop-clause[optseq] new-line
      structured-block  */
 
 #define OACC_LOOP_CLAUSE_MASK						\
 	( (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_COLLAPSE)		\
-	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_REDUCTION))
+	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_REDUCTION) )
 
 static tree
-cp_parser_oacc_loop (cp_parser *parser, cp_token *pragma_tok)
+cp_parser_oacc_loop (cp_parser *parser, cp_token *pragma_tok, char *p_name,
+		     omp_clause_mask mask, tree *cclauses)
 {
-  tree stmt, clauses, block;
-  int save;
+  strcat (p_name, " loop");
+  mask |= OACC_LOOP_CLAUSE_MASK;
 
-  clauses = cp_parser_oacc_all_clauses (parser, OACC_LOOP_CLAUSE_MASK,
-					"#pragma acc loop", pragma_tok);
+  tree clauses = cp_parser_oacc_all_clauses (parser, mask, p_name, pragma_tok,
+					     cclauses == NULL);
+  if (cclauses)
+    {
+      clauses = c_oacc_split_loop_clauses (clauses, cclauses);
+      if (*cclauses)
+	finish_omp_clauses (*cclauses, false);
+      if (clauses)
+	finish_omp_clauses (clauses, false);
+    }
 
-  block = begin_omp_structured_block ();
-  save = cp_parser_begin_omp_structured_block (parser);
-  stmt = cp_parser_omp_for_loop (parser, OACC_LOOP, clauses, NULL);
+  tree block = begin_omp_structured_block ();
+  int save = cp_parser_begin_omp_structured_block (parser);
+  tree stmt = cp_parser_omp_for_loop (parser, OACC_LOOP, clauses, NULL);
   cp_parser_end_omp_structured_block (parser, save);
   add_stmt (finish_omp_structured_block (block));
+
   return stmt;
 }
 
 /* OpenACC 2.0:
+   # pragma acc kernels oacc-kernels-clause[optseq] new-line
+     structured-block
+
+   or
+
    # pragma acc parallel oacc-parallel-clause[optseq] new-line
-     structured-block  */
+     structured-block
+*/
+
+#define OACC_KERNELS_CLAUSE_MASK					\
+	( (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_ASYNC)		\
+	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_COPY)		\
+	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_COPYIN)		\
+	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_COPYOUT)		\
+	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_CREATE)		\
+	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_DEVICEPTR)		\
+	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_IF)			\
+	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_PRESENT)		\
+	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_PRESENT_OR_COPY)	\
+	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_PRESENT_OR_COPYIN)	\
+	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_PRESENT_OR_COPYOUT)	\
+	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_PRESENT_OR_CREATE)	\
+	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_WAIT) )
 
 #define OACC_PARALLEL_CLAUSE_MASK					\
 	( (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_ASYNC)		\
@@ -34379,23 +34374,53 @@ cp_parser_oacc_loop (cp_parser *parser, cp_token *pragma_tok)
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_PRESENT_OR_CREATE)   \
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_REDUCTION)		\
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_VECTOR_LENGTH)       \
-	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_WAIT))
+	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_WAIT) )
 
 static tree
-cp_parser_oacc_parallel (cp_parser *parser, cp_token *pragma_tok)
+cp_parser_oacc_kernels_parallel (cp_parser *parser, cp_token *pragma_tok,
+				 char *p_name)
 {
-  tree stmt, clauses, block;
-  unsigned int save;
+  omp_clause_mask mask;
+  enum tree_code code;
+  switch (pragma_tok->pragma_kind)
+    {
+    case PRAGMA_OACC_KERNELS:
+      strcat (p_name, " kernels");
+      mask = OACC_KERNELS_CLAUSE_MASK;
+      code = OACC_KERNELS;
+      break;
+    case PRAGMA_OACC_PARALLEL:
+      strcat (p_name, " parallel");
+      mask = OACC_PARALLEL_CLAUSE_MASK;
+      code = OACC_PARALLEL;
+      break;
+    default:
+      gcc_unreachable ();
+    }
 
-  clauses = cp_parser_oacc_all_clauses (parser, OACC_PARALLEL_CLAUSE_MASK,
-					 "#pragma acc parallel", pragma_tok);
+  if (cp_lexer_next_token_is (parser->lexer, CPP_NAME))
+    {
+      const char *p
+	= IDENTIFIER_POINTER (cp_lexer_peek_token (parser->lexer)->u.value);
+      if (strcmp (p, "loop") == 0)
+	{
+	  cp_lexer_consume_token (parser->lexer);
+	  mask |= OACC_LOOP_CLAUSE_MASK;
 
-  block = begin_omp_parallel ();
-  save = cp_parser_begin_omp_structured_block (parser);
+	  tree block = begin_omp_parallel ();
+	  tree clauses;
+	  cp_parser_oacc_loop (parser, pragma_tok, p_name, mask, &clauses);
+	  return finish_omp_construct (code, block, clauses);
+	}
+    }
+
+  tree clauses = cp_parser_oacc_all_clauses (parser, mask, p_name, pragma_tok);
+
+  tree block = begin_omp_parallel ();
+  unsigned int save = cp_parser_begin_omp_structured_block (parser);
   cp_parser_statement (parser, NULL_TREE, false, NULL);
   cp_parser_end_omp_structured_block (parser, save);
-  stmt = finish_oacc_parallel (clauses, block);
-  return stmt;
+  return finish_omp_construct (code, block, clauses);
 }
 
 /* OpenACC 2.0:
@@ -35290,13 +35315,13 @@ cp_parser_omp_construct (cp_parser *parser, cp_token *pragma_tok)
       stmt = cp_parser_oacc_enter_exit_data (parser, pragma_tok, false);
       break;
     case PRAGMA_OACC_KERNELS:
-      stmt = cp_parser_oacc_kernels (parser, pragma_tok);
+    case PRAGMA_OACC_PARALLEL:
+      strcpy (p_name, "#pragma acc");
+      stmt = cp_parser_oacc_kernels_parallel (parser, pragma_tok, p_name);
       break;
     case PRAGMA_OACC_LOOP:
-      stmt = cp_parser_oacc_loop (parser, pragma_tok);
-      break;
-    case PRAGMA_OACC_PARALLEL:
-      stmt = cp_parser_oacc_parallel (parser, pragma_tok);
+      strcpy (p_name, "#pragma acc");
+      stmt = cp_parser_oacc_loop (parser, pragma_tok, p_name, mask, NULL);
       break;
     case PRAGMA_OACC_UPDATE:
       stmt = cp_parser_oacc_update (parser, pragma_tok);
diff --git a/gcc/cp/semantics.c b/gcc/cp/semantics.c
index 11315d9..f9e86d0 100644
--- a/gcc/cp/semantics.c
+++ b/gcc/cp/semantics.c
@@ -7106,69 +7106,49 @@ finish_omp_structured_block (tree block)
   return do_poplevel (block);
 }
 
-/* Generate OACC_DATA, with CLAUSES and BLOCK as its compound
-   statement.  LOC is the location of the OACC_DATA.  */
+/* Similarly, except force the retention of the BLOCK.  */
 
 tree
-finish_oacc_data (tree clauses, tree block)
+begin_omp_parallel (void)
 {
-  tree stmt;
-
-  block = finish_omp_structured_block (block);
-
-  stmt = make_node (OACC_DATA);
-  TREE_TYPE (stmt) = void_type_node;
-  OACC_DATA_CLAUSES (stmt) = clauses;
-  OACC_DATA_BODY (stmt) = block;
-
-  return add_stmt (stmt);
+  keep_next_level (true);
+  return begin_omp_structured_block ();
 }
 
-/* Generate OACC_KERNELS, with CLAUSES and BLOCK as its compound
-   statement.  LOC is the location of the OACC_KERNELS.  */
+/* Generate OACC_DATA, with CLAUSES and BLOCK as its compound
+   statement.  */
 
 tree
-finish_oacc_kernels (tree clauses, tree block)
+finish_oacc_data (tree clauses, tree block)
 {
   tree stmt;
 
   block = finish_omp_structured_block (block);
 
-  stmt = make_node (OACC_KERNELS);
+  stmt = make_node (OACC_DATA);
   TREE_TYPE (stmt) = void_type_node;
-  OACC_KERNELS_CLAUSES (stmt) = clauses;
-  OACC_KERNELS_BODY (stmt) = block;
+  OACC_DATA_CLAUSES (stmt) = clauses;
+  OACC_DATA_BODY (stmt) = block;
 
   return add_stmt (stmt);
 }
 
-/* Generate OACC_PARALLEL, with CLAUSES and BLOCK as its compound
-   statement.  LOC is the location of the OACC_PARALLEL.  */
+/* Generate OMP construct CODE, with BODY and CLAUSES as its compound
+   statement.  */
 
 tree
-finish_oacc_parallel (tree clauses, tree block)
+finish_omp_construct (enum tree_code code, tree body, tree clauses)
 {
-  tree stmt;
-
-  block = finish_omp_structured_block (block);
+  body = finish_omp_structured_block (body);
 
-  stmt = make_node (OACC_PARALLEL);
+  tree stmt = make_node (code);
   TREE_TYPE (stmt) = void_type_node;
-  OACC_PARALLEL_CLAUSES (stmt) = clauses;
-  OACC_PARALLEL_BODY (stmt) = block;
+  OMP_BODY (stmt) = body;
+  OMP_CLAUSES (stmt) = clauses;
 
   return add_stmt (stmt);
 }
 
-/* Similarly, except force the retention of the BLOCK.  */
-
-tree
-begin_omp_parallel (void)
-{
-  keep_next_level (true);
-  return begin_omp_structured_block ();
-}
-
 tree
 finish_omp_parallel (tree clauses, tree body)
 {