Update OpenACC data clause semantics to the 2.5 behavior

	gcc/c-family/
	* c-pragma.h (enum pragma_omp_clause): Add
	PRAGMA_OACC_CLAUSE_{FINALIZE,IF_PRESENT}. Remove
	PRAGMA_OACC_CLAUSE_PRESENT_OR_{COPY,COPYIN,COPYOUT,CREATE}.

	gcc/c/
	* c-parser.c (c_parser_omp_clause_name): Add support for finalize
	and if_present. Make present_or_{copy,copyin,copyout,create} aliases
	to their non-present_or_* counterparts. Make 'self' an alias to
	PRAGMA_OACC_CLAUSE_HOST.
	(c_parser_oacc_data_clause): Update GOMP mappings for
	PRAGMA_OACC_CLAUSE_{COPY,COPYIN,COPYOUT,CREATE,DELETE}. Remove
	PRAGMA_OACC_CLAUSE_{SELF,PRESENT_OR_*}.
	(c_parser_oacc_all_clauses): Handle finalize and if_present clauses.
	Remove support for present_or_* clauses.
	(OACC_KERNELS_CLAUSE_MASK): Remove PRESENT_OR_* clauses.
	(OACC_PARALLEL_CLAUSE_MASK): Likewise.
	(OACC_DECLARE_CLAUSE_MASK): Likewise.
	(OACC_DATA_CLAUSE_MASK): Likewise.
	(OACC_ENTER_DATA_CLAUSE_MASK): Remove PRESENT_OR_* clauses.
	(OACC_EXIT_DATA_CLAUSE_MASK): Add FINALIZE clause.
	(OACC_UPDATE_CLAUSE_MASK): Remove SELF, add IF_PRESENT.
	(c_parser_oacc_declare): Remove PRESENT_OR_* clauses.
	* c-typeck.c (c_finish_omp_clauses): Handle IF_PRESENT and FINALIZE.

	gcc/cp/
	* parser.c (cp_parser_omp_clause_name): Add support for finalize
	and if_present. Make present_or_{copy,copyin,copyout,create} aliases
	to their non-present_or_* counterparts. Make 'self' an alias to
	PRAGMA_OACC_CLAUSE_HOST.
	(cp_parser_oacc_data_clause): Update GOMP mappings for
	PRAGMA_OACC_CLAUSE_{COPY,COPYIN,COPYOUT,CREATE,DELETE}. Remove
	PRAGMA_OACC_CLAUSE_{SELF,PRESENT_OR_*}.
	(cp_parser_oacc_all_clauses): Handle finalize and if_present clauses.
	Remove support for present_or_* clauses.
	(OACC_KERNELS_CLAUSE_MASK): Remove PRESENT_OR_* clauses.
	(OACC_PARALLEL_CLAUSE_MASK): Likewise.
	(OACC_DECLARE_CLAUSE_MASK): Likewise.
	(OACC_DATA_CLAUSE_MASK): Likewise.
	(OACC_ENTER_DATA_CLAUSE_MASK): Remove PRESENT_OR_* clauses.
	(OACC_EXIT_DATA_CLAUSE_MASK): Add FINALIZE clause.
	(OACC_UPDATE_CLAUSE_MASK): Remove SELF, add IF_PRESENT.
	(cp_parser_oacc_declare): Remove PRESENT_OR_* clauses.
	* pt.c (tsubst_omp_clauses): Handle IF_PRESENT and FINALIZE.
	* semantics.c (finish_omp_clauses): Handle IF_PRESENT and FINALIZE.

	gcc/fortran/
	* gfortran.h (gfc_omp_clauses): Add unsigned if_present, finalize
	bitfields.
	* openmp.c (enum omp_mask2): Remove OMP_CLAUSE_PRESENT_OR_*. Add
	OMP_CLAUSE_{IF_PRESENT,FINALIZE}.
	(gfc_match_omp_clauses): Update handling of copy, copyin, copyout,
	create, deviceptr, present_of_*. Add support for finalize and
	if_present.
	(OACC_PARALLEL_CLAUSES): Remove PRESENT_OR_* clauses.
	(OACC_KERNELS_CLAUSES): Likewise.
	(OACC_DATA_CLAUSES): Likewise.
	(OACC_DECLARE_CLAUSES): Likewise.
	(OACC_UPDATE_CLAUSES): Add IF_PRESENT clause.
	(OACC_ENTER_DATA_CLAUSES): Remove PRESENT_OR_* clauses.
	(OACC_EXIT_DATA_CLAUSES): Add FINALIZE clause.
	(gfc_match_oacc_declare): Update to OpenACC 2.5 semantics.
	* trans-openmp.c (gfc_trans_omp_clauses): Add support for IF_PRESENT
	and FINALIZE.

	gcc/
	* gimplify.c (gimplify_scan_omp_clauses): Add support for
	OMP_CLAUSE_{IF_PRESENT,FINALIZE}.
	(gimplify_adjust_omp_clauses): Likewise.
	(gimplify_oacc_declare_1): Add support for GOMP_MAP_RELEASE, remove
	support for GOMP_MAP_FORCE_{ALLOC,TO,FROM,TOFROM}.
	(gimplify_omp_target_update): Update handling of acc update and
	enter/exit data.
	* omp-low.c (install_var_field): Remove unused parameter
	base_pointers_restrict.
	(scan_sharing_clauses): Remove base_pointers_restrict parameter.
	Update call to install_var_field. Handle OMP_CLAUSE_{IF_PRESENT,
	FINALIZE}
	(omp_target_base_pointers_restrict_p): Delete.
	(scan_omp_target): Update call to scan_sharing_clauses.
	* tree-core.h (enum omp_clause_code): Add OMP_CLAUSE_{IF_PRESENT,
	FINALIZE}.
	* tree-nested.c (convert_nonlocal_omp_clauses): Handle
	OMP_CLAUSE_{IF_PRESENT,FINALIZE}.
	(convert_local_omp_clauses): Likewise.
	* tree-pretty-print.c (dump_omp_clause): Likewise.
	* tree.c (omp_clause_num_ops): Add entries for 	OMP_CLAUSE_{IF_PRESENT,
	FINALIZE}.
	(omp_clause_code_name): Likewise.

	gcc/testsuite/
	* c-c++-common/goacc/declare-1.c: Update test case to utilize OpenACC
	2.5 data clause semantics.
	* c-c++-common/goacc/declare-2.c: Likewise.
	* c-c++-common/goacc/default-4.c: Likewise.
	* c-c++-common/goacc/finalize-1.c: New test.
	* c-c++-common/goacc/kernels-alias-2.c: Update test case to utilize
	OpenACC 2.5 data clause semantics.
	* c-c++-common/goacc/kernels-alias.c: Likewise.
	* c-c++-common/goacc/routine-5.c: Likewise.
	* c-c++-common/goacc/update-if_present-1.c: New test.
	* c-c++-common/goacc/update-if_present-2.c: New test.
	* g++.dg/goacc/template.C: Update test case to utilize OpenACC
	2.5 data clause semantics.
	* gfortran.dg/goacc/combined-directives.f90: Likewise.
	* gfortran.dg/goacc/data-tree.f95: Likewise.
	* gfortran.dg/goacc/declare-2.f95: Likewise.
	* gfortran.dg/goacc/default-4.f: Likewise.
	* gfortran.dg/goacc/enter-exit-data.f95: Likewise.
	* gfortran.dg/goacc/finalize-1.f: New test.
	* gfortran.dg/goacc/kernels-alias-2.f95: Update test case to utilize
	OpenACC 2.5 data clause semantics.
	* gfortran.dg/goacc/kernels-alias.f95: Likewise.
	* gfortran.dg/goacc/kernels-tree.f95: Likewise.
	* gfortran.dg/goacc/nested-function-1.f90: Likewise.
	* gfortran.dg/goacc/parallel-tree.f95: Likewise.
	* gfortran.dg/goacc/reduction-promotions.f90: Likewise.
	* gfortran.dg/goacc/update-if_present-1.f90: New test.
	* gfortran.dg/goacc/update-if_present-2.f90: New test.

	libgomp/
	* libgomp.h (struct splay_tree_key_s): Add dynamic_refcount member.
	(gomp_acc_remove_pointer): Update declaration.
	(gomp_acc_declare_allocate): Declare.
	(gomp_remove_var): Declare.
	* libgomp.map (OACC_2.5): Define.
	* oacc-mem.c (acc_map_data): Update refcount.
	(acc_unmap_data): Likewise.
	(present_create_copy): Likewise.
	(acc_create): Add FLAG_PRESENT when calling present_create_copy.
	(acc_copyin): Likewise.
	(FLAG_FINALIZE): Define.
	(delete_copyout): Update dynamic refcounts, add support for FINALIZE.
	(acc_delete_finalize): New function.
	(acc_delete_finalize_async): New function.
	(acc_copyout_finalize): New function.
	(acc_copyout_finalize_async): New function.
	(gomp_acc_insert_pointer): Update refcounts.
	(gomp_acc_remove_pointer): Return if data is not present on the
	accelerator.
	* oacc-parallel.c (find_pset): Rename to find_pointer.
	(find_pointer): Add support for GOMP_MAP_POINTER.
	(handle_ftn_pointers): New function.
	(GOACC_parallel_keyed): Update refcounts of variables.
	(GOACC_enter_exit_data): Add support for finalized data mappings.
	Add support for GOMP_MAP_{TO,ALLOC,RELESE,FROM}. Update handling
	of fortran arrays.
	(GOACC_update): Add support for GOMP_MAP_{ALWAYS_POINTER,TO,FROM}.
	(GOACC_declare): Add support for GOMP_MAP_RELEASE, remove support
	for GOMP_MAP_FORCE_FROM.
	* openacc.f90 (module openacc_internal): Add
	acc_copyout_finalize_{32_h,64_h,array_h,_l}, and
	acc_delete_finalize_{32_h,64_h,array_h,_l}. Add interfaces for
	acc_copyout_finalize and acc_delete_finalize.
	(acc_copyout_finalize_32_h): New subroutine.
	(acc_copyout_finalize_64_h): New subroutine.
	(acc_copyout_finalize_array_h): New subroutine.
	(acc_delete_finalize_32_h): New subroutine.
	(acc_delete_finalize_64_h): New subroutine.
	(acc_delete_finalize_array_h): New subroutine.
	* openacc.h (acc_copyout_finalize): Declare.
	(acc_copyout_finalize_async): Declare.
	(acc_delete_finalize): Declare.
	(acc_delete_finalize_async): Declare.
	* openacc_lib.h (acc_copyout_finalize): New interface.
	(acc_delete_finalize): New interface.
	* target.c (gomp_map_vars): Update dynamic_refcount.
	(gomp_remove_var): New function.
	(gomp_unmap_vars): Use it.
	(gomp_unload_image_from_device): Likewise.
	* testsuite/libgomp.oacc-c-c++-common/data-already-1.c: Update test
	case to utilize OpenACC 2.5 data clause semantics.
	* testsuite/libgomp.oacc-c-c++-common/data-already-2.c: Likewise.
	* testsuite/libgomp.oacc-c-c++-common/data-already-3.c: Likewise.
	* testsuite/libgomp.oacc-c-c++-common/data-already-4.c: Likewise.
	* testsuite/libgomp.oacc-c-c++-common/data-already-5.c: Likewise.
	* testsuite/libgomp.oacc-c-c++-common/data-already-6.c: Likewise.
	* testsuite/libgomp.oacc-c-c++-common/data-already-7.c: Likewise.
	* testsuite/libgomp.oacc-c-c++-common/data-already-8.c: Likewise.
	* testsuite/libgomp.oacc-c-c++-common/lib-16.c: Likewise.
	* testsuite/libgomp.oacc-c-c++-common/lib-25.c: Likewise.
	* testsuite/libgomp.oacc-c-c++-common/lib-32.c: Likewise.
	* testsuite/libgomp.oacc-c-c++-common/lib-83.c: Likewise.
	* testsuite/libgomp.oacc-fortran/data-5.f90: New test.
	* testsuite/libgomp.oacc-fortran/data-already-1.f: Update test case to
	utilize OpenACC 2.5 data clause semantics.
	* testsuite/libgomp.oacc-fortran/data-already-2.f: Likewise.
	* testsuite/libgomp.oacc-fortran/data-already-3.f: Likewise.
	* testsuite/libgomp.oacc-fortran/data-already-4.f: Likewise.
	* testsuite/libgomp.oacc-fortran/data-already-5.f: Likewise.
	* testsuite/libgomp.oacc-fortran/data-already-6.f: Likewise.
	* testsuite/libgomp.oacc-fortran/data-already-7.f: Likewise.
	* testsuite/libgomp.oacc-fortran/data-already-8.f: Likewise.
	* testsuite/libgomp.oacc-fortran/lib-32-1.f: Likewise.
	* testsuite/libgomp.oacc-fortran/lib-32-2.f: Likewise.

Co-Authored-By: Cesar Philippidis <cesar@codesourcery.com>
Co-Authored-By: Thomas Schwinge <thomas@codesourcery.com>

From-SVN: r261813

45 files changed, 717 insertions, 466 deletions

diff --git a/gcc/ChangeLog b/gcc/ChangeLog
index 7b399f1..37fc7da 100644
--- a/gcc/ChangeLog
+++ b/gcc/ChangeLog
@@ -1,3 +1,31 @@
+2018-06-20  Chung-Lin Tang <cltang@codesourcery.com>
+	    Thomas Schwinge <thomas@codesourcery.com>
+	    Cesar Philippidis  <cesar@codesourcery.com>
+
+	* gimplify.c (gimplify_scan_omp_clauses): Add support for
+	OMP_CLAUSE_{IF_PRESENT,FINALIZE}.
+	(gimplify_adjust_omp_clauses): Likewise.
+	(gimplify_oacc_declare_1): Add support for GOMP_MAP_RELEASE, remove
+	support for GOMP_MAP_FORCE_{ALLOC,TO,FROM,TOFROM}.
+	(gimplify_omp_target_update): Update handling of acc update and
+	enter/exit data.
+	* omp-low.c (install_var_field): Remove unused parameter
+	base_pointers_restrict.
+	(scan_sharing_clauses): Remove base_pointers_restrict parameter.
+	Update call to install_var_field. Handle OMP_CLAUSE_{IF_PRESENT,
+	FINALIZE}
+	(omp_target_base_pointers_restrict_p): Delete.
+	(scan_omp_target): Update call to scan_sharing_clauses.
+	* tree-core.h (enum omp_clause_code): Add OMP_CLAUSE_{IF_PRESENT,
+	FINALIZE}.
+	* tree-nested.c (convert_nonlocal_omp_clauses): Handle
+	OMP_CLAUSE_{IF_PRESENT,FINALIZE}.
+	(convert_local_omp_clauses): Likewise.
+	* tree-pretty-print.c (dump_omp_clause): Likewise.
+	* tree.c (omp_clause_num_ops): Add entries for 	OMP_CLAUSE_{IF_PRESENT,
+	FINALIZE}.
+	(omp_clause_code_name): Likewise.
+
 2018-06-20  Jakub Jelinek  <jakub@redhat.com>
 
 	PR debug/86194
diff --git a/gcc/c-family/ChangeLog b/gcc/c-family/ChangeLog
index 6b974da..fc7f7de 100644
--- a/gcc/c-family/ChangeLog
+++ b/gcc/c-family/ChangeLog
@@ -1,3 +1,11 @@
+2018-06-20  Chung-Lin Tang <cltang@codesourcery.com>
+	    Thomas Schwinge <thomas@codesourcery.com>
+	    Cesar Philippidis  <cesar@codesourcery.com>
+
+	* c-pragma.h (enum pragma_omp_clause): Add
+	PRAGMA_OACC_CLAUSE_{FINALIZE,IF_PRESENT}. Remove
+	PRAGMA_OACC_CLAUSE_PRESENT_OR_{COPY,COPYIN,COPYOUT,CREATE}.
+
 2018-06-20  Jakub Jelinek  <jakub@redhat.com>
 
 	PR c++/86210
diff --git a/gcc/c-family/c-pragma.h b/gcc/c-family/c-pragma.h
index c70380c..b322547 100644
--- a/gcc/c-family/c-pragma.h
+++ b/gcc/c-family/c-pragma.h
@@ -138,16 +138,13 @@ enum pragma_omp_clause {
   PRAGMA_OACC_CLAUSE_DELETE,
   PRAGMA_OACC_CLAUSE_DEVICEPTR,
   PRAGMA_OACC_CLAUSE_DEVICE_RESIDENT,
+  PRAGMA_OACC_CLAUSE_FINALIZE,
   PRAGMA_OACC_CLAUSE_GANG,
   PRAGMA_OACC_CLAUSE_HOST,
   PRAGMA_OACC_CLAUSE_INDEPENDENT,
   PRAGMA_OACC_CLAUSE_NUM_GANGS,
   PRAGMA_OACC_CLAUSE_NUM_WORKERS,
   PRAGMA_OACC_CLAUSE_PRESENT,
-  PRAGMA_OACC_CLAUSE_PRESENT_OR_COPY,
-  PRAGMA_OACC_CLAUSE_PRESENT_OR_COPYIN,
-  PRAGMA_OACC_CLAUSE_PRESENT_OR_COPYOUT,
-  PRAGMA_OACC_CLAUSE_PRESENT_OR_CREATE,
   PRAGMA_OACC_CLAUSE_SELF,
   PRAGMA_OACC_CLAUSE_SEQ,
   PRAGMA_OACC_CLAUSE_TILE,
@@ -156,6 +153,7 @@ enum pragma_omp_clause {
   PRAGMA_OACC_CLAUSE_VECTOR_LENGTH,
   PRAGMA_OACC_CLAUSE_WAIT,
   PRAGMA_OACC_CLAUSE_WORKER,
+  PRAGMA_OACC_CLAUSE_IF_PRESENT,
   PRAGMA_OACC_CLAUSE_COLLAPSE = PRAGMA_OMP_CLAUSE_COLLAPSE,
   PRAGMA_OACC_CLAUSE_COPYIN = PRAGMA_OMP_CLAUSE_COPYIN,
   PRAGMA_OACC_CLAUSE_DEVICE = PRAGMA_OMP_CLAUSE_DEVICE,
diff --git a/gcc/c/ChangeLog b/gcc/c/ChangeLog
index f1ff69e8..1ba1173 100644
--- a/gcc/c/ChangeLog
+++ b/gcc/c/ChangeLog
@@ -1,3 +1,26 @@
+2018-06-20  Chung-Lin Tang <cltang@codesourcery.com>
+	    Thomas Schwinge <thomas@codesourcery.com>
+	    Cesar Philippidis  <cesar@codesourcery.com>
+
+	* c-parser.c (c_parser_omp_clause_name): Add support for finalize
+	and if_present. Make present_or_{copy,copyin,copyout,create} aliases
+	to their non-present_or_* counterparts. Make 'self' an alias to
+	PRAGMA_OACC_CLAUSE_HOST.
+	(c_parser_oacc_data_clause): Update GOMP mappings for
+	PRAGMA_OACC_CLAUSE_{COPY,COPYIN,COPYOUT,CREATE,DELETE}. Remove
+	PRAGMA_OACC_CLAUSE_{SELF,PRESENT_OR_*}.
+	(c_parser_oacc_all_clauses): Handle finalize and if_present clauses.
+	Remove support for present_or_* clauses.
+	(OACC_KERNELS_CLAUSE_MASK): Remove PRESENT_OR_* clauses.
+	(OACC_PARALLEL_CLAUSE_MASK): Likewise.
+	(OACC_DECLARE_CLAUSE_MASK): Likewise.
+	(OACC_DATA_CLAUSE_MASK): Likewise.
+	(OACC_ENTER_DATA_CLAUSE_MASK): Remove PRESENT_OR_* clauses.
+	(OACC_EXIT_DATA_CLAUSE_MASK): Add FINALIZE clause.
+	(OACC_UPDATE_CLAUSE_MASK): Remove SELF, add IF_PRESENT.
+	(c_parser_oacc_declare): Remove PRESENT_OR_* clauses.
+	* c-typeck.c (c_finish_omp_clauses): Handle IF_PRESENT and FINALIZE.
+
 2018-06-16  Kugan Vivekanandarajah  <kuganv@linaro.org>
 
 	* c-typeck.c (build_unary_op): Handle ABSU_EXPR;
diff --git a/gcc/c/c-parser.c b/gcc/c/c-parser.c
index 6b41a61..7a92628 100644
--- a/gcc/c/c-parser.c
+++ b/gcc/c/c-parser.c
@@ -11259,6 +11259,8 @@ c_parser_omp_clause_name (c_parser *parser)
 	case 'f':
 	  if (!strcmp ("final", p))
 	    result = PRAGMA_OMP_CLAUSE_FINAL;
+	  else if (!strcmp ("finalize", p))
+	    result = PRAGMA_OACC_CLAUSE_FINALIZE;
 	  else if (!strcmp ("firstprivate", p))
 	    result = PRAGMA_OMP_CLAUSE_FIRSTPRIVATE;
 	  else if (!strcmp ("from", p))
@@ -11277,7 +11279,9 @@ c_parser_omp_clause_name (c_parser *parser)
 	    result = PRAGMA_OACC_CLAUSE_HOST;
 	  break;
 	case 'i':
-	  if (!strcmp ("inbranch", p))
+	  if (!strcmp ("if_present", p))
+	    result = PRAGMA_OACC_CLAUSE_IF_PRESENT;
+	  else if (!strcmp ("inbranch", p))
 	    result = PRAGMA_OMP_CLAUSE_INBRANCH;
 	  else if (!strcmp ("independent", p))
 	    result = PRAGMA_OACC_CLAUSE_INDEPENDENT;
@@ -11325,18 +11329,20 @@ c_parser_omp_clause_name (c_parser *parser)
 	    result = PRAGMA_OMP_CLAUSE_PARALLEL;
 	  else if (!strcmp ("present", p))
 	    result = PRAGMA_OACC_CLAUSE_PRESENT;
+	  /* As of OpenACC 2.5, these are now aliases of the non-present_or
+	     clauses.  */
 	  else if (!strcmp ("present_or_copy", p)
 		   || !strcmp ("pcopy", p))
-	    result = PRAGMA_OACC_CLAUSE_PRESENT_OR_COPY;
+	    result = PRAGMA_OACC_CLAUSE_COPY;
 	  else if (!strcmp ("present_or_copyin", p)
 		   || !strcmp ("pcopyin", p))
-	    result = PRAGMA_OACC_CLAUSE_PRESENT_OR_COPYIN;
+	    result = PRAGMA_OACC_CLAUSE_COPYIN;
 	  else if (!strcmp ("present_or_copyout", p)
 		   || !strcmp ("pcopyout", p))
-	    result = PRAGMA_OACC_CLAUSE_PRESENT_OR_COPYOUT;
+	    result = PRAGMA_OACC_CLAUSE_COPYOUT;
 	  else if (!strcmp ("present_or_create", p)
 		   || !strcmp ("pcreate", p))
-	    result = PRAGMA_OACC_CLAUSE_PRESENT_OR_CREATE;
+	    result = PRAGMA_OACC_CLAUSE_CREATE;
 	  else if (!strcmp ("priority", p))
 	    result = PRAGMA_OMP_CLAUSE_PRIORITY;
 	  else if (!strcmp ("private", p))
@@ -11355,6 +11361,8 @@ c_parser_omp_clause_name (c_parser *parser)
 	    result = PRAGMA_OMP_CLAUSE_SCHEDULE;
 	  else if (!strcmp ("sections", p))
 	    result = PRAGMA_OMP_CLAUSE_SECTIONS;
+	  else if (!strcmp ("self", p)) /* "self" is a synonym for "host".  */
+	    result = PRAGMA_OACC_CLAUSE_HOST;
 	  else if (!strcmp ("seq", p))
 	    result = PRAGMA_OACC_CLAUSE_SEQ;
 	  else if (!strcmp ("shared", p))
@@ -11363,8 +11371,6 @@ c_parser_omp_clause_name (c_parser *parser)
 	    result = PRAGMA_OMP_CLAUSE_SIMD;
 	  else if (!strcmp ("simdlen", p))
 	    result = PRAGMA_OMP_CLAUSE_SIMDLEN;
-	  else if (!strcmp ("self", p))
-	    result = PRAGMA_OACC_CLAUSE_SELF;
 	  break;
 	case 't':
 	  if (!strcmp ("taskgroup", p))
@@ -11646,15 +11652,7 @@ c_parser_omp_var_list_parens (c_parser *parser, enum omp_clause_code kind,
    copyout ( variable-list )
    create ( variable-list )
    delete ( variable-list )
-   present ( variable-list )
-   present_or_copy ( variable-list )
-     pcopy ( variable-list )
-   present_or_copyin ( variable-list )
-     pcopyin ( variable-list )
-   present_or_copyout ( variable-list )
-     pcopyout ( variable-list )
-   present_or_create ( variable-list )
-     pcreate ( variable-list ) */
+   present ( variable-list ) */
 
 static tree
 c_parser_oacc_data_clause (c_parser *parser, pragma_omp_clause c_kind,
@@ -11664,19 +11662,19 @@ c_parser_oacc_data_clause (c_parser *parser, pragma_omp_clause c_kind,
   switch (c_kind)
     {
     case PRAGMA_OACC_CLAUSE_COPY:
-      kind = GOMP_MAP_FORCE_TOFROM;
+      kind = GOMP_MAP_TOFROM;
       break;
     case PRAGMA_OACC_CLAUSE_COPYIN:
-      kind = GOMP_MAP_FORCE_TO;
+      kind = GOMP_MAP_TO;
       break;
     case PRAGMA_OACC_CLAUSE_COPYOUT:
-      kind = GOMP_MAP_FORCE_FROM;
+      kind = GOMP_MAP_FROM;
       break;
     case PRAGMA_OACC_CLAUSE_CREATE:
-      kind = GOMP_MAP_FORCE_ALLOC;
+      kind = GOMP_MAP_ALLOC;
       break;
     case PRAGMA_OACC_CLAUSE_DELETE:
-      kind = GOMP_MAP_DELETE;
+      kind = GOMP_MAP_RELEASE;
       break;
     case PRAGMA_OACC_CLAUSE_DEVICE:
       kind = GOMP_MAP_FORCE_TO;
@@ -11685,7 +11683,6 @@ c_parser_oacc_data_clause (c_parser *parser, pragma_omp_clause c_kind,
       kind = GOMP_MAP_DEVICE_RESIDENT;
       break;
     case PRAGMA_OACC_CLAUSE_HOST:
-    case PRAGMA_OACC_CLAUSE_SELF:
       kind = GOMP_MAP_FORCE_FROM;
       break;
     case PRAGMA_OACC_CLAUSE_LINK:
@@ -11694,18 +11691,6 @@ c_parser_oacc_data_clause (c_parser *parser, pragma_omp_clause c_kind,
     case PRAGMA_OACC_CLAUSE_PRESENT:
       kind = GOMP_MAP_FORCE_PRESENT;
       break;
-    case PRAGMA_OACC_CLAUSE_PRESENT_OR_COPY:
-      kind = GOMP_MAP_TOFROM;
-      break;
-    case PRAGMA_OACC_CLAUSE_PRESENT_OR_COPYIN:
-      kind = GOMP_MAP_TO;
-      break;
-    case PRAGMA_OACC_CLAUSE_PRESENT_OR_COPYOUT:
-      kind = GOMP_MAP_FROM;
-      break;
-    case PRAGMA_OACC_CLAUSE_PRESENT_OR_CREATE:
-      kind = GOMP_MAP_ALLOC;
-      break;
     default:
       gcc_unreachable ();
     }
@@ -12597,8 +12582,9 @@ c_parser_oacc_shape_clause (c_parser *parser, omp_clause_code kind,
   return list;
 }
 
-/* OpenACC:
+/* OpenACC 2.5:
    auto
+   finalize
    independent
    nohost
    seq */
@@ -13955,6 +13941,11 @@ c_parser_oacc_all_clauses (c_parser *parser, omp_clause_mask mask,
 	  clauses = c_parser_oacc_data_clause (parser, c_kind, clauses);
 	  c_name = "device_resident";
 	  break;
+	case PRAGMA_OACC_CLAUSE_FINALIZE:
+	  clauses = c_parser_oacc_simple_clause (parser, OMP_CLAUSE_FINALIZE,
+						 clauses);
+	  c_name = "finalize";
+	  break;
 	case PRAGMA_OACC_CLAUSE_FIRSTPRIVATE:
 	  clauses = c_parser_omp_clause_firstprivate (parser, clauses);
 	  c_name = "firstprivate";
@@ -13972,6 +13963,11 @@ c_parser_oacc_all_clauses (c_parser *parser, omp_clause_mask mask,
 	  clauses = c_parser_omp_clause_if (parser, clauses, false);
 	  c_name = "if";
 	  break;
+	case PRAGMA_OACC_CLAUSE_IF_PRESENT:
+	  clauses = c_parser_oacc_simple_clause (parser, OMP_CLAUSE_IF_PRESENT,
+						 clauses);
+	  c_name = "if_present";
+	  break;
 	case PRAGMA_OACC_CLAUSE_INDEPENDENT:
 	  clauses = c_parser_oacc_simple_clause (parser, OMP_CLAUSE_INDEPENDENT,
 						clauses);
@@ -13997,22 +13993,6 @@ c_parser_oacc_all_clauses (c_parser *parser, omp_clause_mask mask,
 	  clauses = c_parser_oacc_data_clause (parser, c_kind, clauses);
 	  c_name = "present";
 	  break;
-	case PRAGMA_OACC_CLAUSE_PRESENT_OR_COPY:
-	  clauses = c_parser_oacc_data_clause (parser, c_kind, clauses);
-	  c_name = "present_or_copy";
-	  break;
-	case PRAGMA_OACC_CLAUSE_PRESENT_OR_COPYIN:
-	  clauses = c_parser_oacc_data_clause (parser, c_kind, clauses);
-	  c_name = "present_or_copyin";
-	  break;
-	case PRAGMA_OACC_CLAUSE_PRESENT_OR_COPYOUT:
-	  clauses = c_parser_oacc_data_clause (parser, c_kind, clauses);
-	  c_name = "present_or_copyout";
-	  break;
-	case PRAGMA_OACC_CLAUSE_PRESENT_OR_CREATE:
-	  clauses = c_parser_oacc_data_clause (parser, c_kind, clauses);
-	  c_name = "present_or_create";
-	  break;
 	case PRAGMA_OACC_CLAUSE_PRIVATE:
 	  clauses = c_parser_omp_clause_private (parser, clauses);
 	  c_name = "private";
@@ -14021,10 +14001,6 @@ c_parser_oacc_all_clauses (c_parser *parser, omp_clause_mask mask,
 	  clauses = c_parser_omp_clause_reduction (parser, clauses);
 	  c_name = "reduction";
 	  break;
-	case PRAGMA_OACC_CLAUSE_SELF:
-	  clauses = c_parser_oacc_data_clause (parser, c_kind, clauses);
-	  c_name = "self";
-	  break;
 	case PRAGMA_OACC_CLAUSE_SEQ:
 	  clauses = c_parser_oacc_simple_clause (parser, OMP_CLAUSE_SEQ,
 						clauses);
@@ -14417,11 +14393,7 @@ c_parser_oacc_cache (location_t loc, c_parser *parser)
 	| (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_PRESENT))
 
 static tree
 c_parser_oacc_data (location_t loc, c_parser *parser, bool *if_p)
@@ -14451,11 +14423,7 @@ c_parser_oacc_data (location_t loc, c_parser *parser, bool *if_p)
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_DEVICEPTR)		\
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_DEVICE_RESIDENT)	\
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_LINK)		\
-	| (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_PRESENT))
 
 static void
 c_parser_oacc_declare (c_parser *parser)
@@ -14490,8 +14458,8 @@ c_parser_oacc_declare (c_parser *parser)
       switch (OMP_CLAUSE_MAP_KIND (t))
 	{
 	case GOMP_MAP_FIRSTPRIVATE_POINTER:
-	case GOMP_MAP_FORCE_ALLOC:
-	case GOMP_MAP_FORCE_TO:
+	case GOMP_MAP_ALLOC:
+	case GOMP_MAP_TO:
 	case GOMP_MAP_FORCE_DEVICEPTR:
 	case GOMP_MAP_DEVICE_RESIDENT:
 	  break;
@@ -14604,8 +14572,6 @@ c_parser_oacc_declare (c_parser *parser)
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_ASYNC)		\
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_COPYIN)		\
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_CREATE)		\
-	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_PRESENT_OR_COPYIN)	\
-	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_PRESENT_OR_CREATE)	\
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_WAIT) )
 
 #define OACC_EXIT_DATA_CLAUSE_MASK					\
@@ -14613,6 +14579,7 @@ c_parser_oacc_declare (c_parser *parser)
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_ASYNC)		\
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_COPYOUT)		\
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_DELETE) 		\
+	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_FINALIZE) 		\
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_WAIT) )
 
 static void
@@ -14756,10 +14723,6 @@ c_parser_oacc_loop (location_t loc, c_parser *parser, char *p_name,
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_NUM_GANGS)		\
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_NUM_WORKERS)		\
 	| (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_VECTOR_LENGTH)	\
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_WAIT) )
 
@@ -14777,10 +14740,6 @@ c_parser_oacc_loop (location_t loc, c_parser *parser, char *p_name,
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_NUM_GANGS)		\
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_NUM_WORKERS)		\
 	| (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_REDUCTION)		\
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_VECTOR_LENGTH)	\
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_WAIT) )
@@ -15008,7 +14967,7 @@ c_finish_oacc_routine (struct oacc_routine_data *data, tree fndecl,
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_DEVICE)		\
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_HOST)		\
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_IF)			\
-	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_SELF)		\
+	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_IF_PRESENT)		\
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_WAIT) )
 
 static void
diff --git a/gcc/c/c-typeck.c b/gcc/c/c-typeck.c
index aa70b23..90ae306 100644
--- a/gcc/c/c-typeck.c
+++ b/gcc/c/c-typeck.c
@@ -13897,6 +13897,8 @@ c_finish_omp_clauses (tree clauses, enum c_omp_region_type ort)
 	case OMP_CLAUSE_WORKER:
 	case OMP_CLAUSE_VECTOR:
 	case OMP_CLAUSE_TILE:
+	case OMP_CLAUSE_IF_PRESENT:
+	case OMP_CLAUSE_FINALIZE:
 	  pc = &OMP_CLAUSE_CHAIN (c);
 	  continue;
 
diff --git a/gcc/cp/ChangeLog b/gcc/cp/ChangeLog
index 72cd388..56cecfa 100644
--- a/gcc/cp/ChangeLog
+++ b/gcc/cp/ChangeLog
@@ -1,3 +1,27 @@
+2018-06-20  Chung-Lin Tang <cltang@codesourcery.com>
+	    Thomas Schwinge <thomas@codesourcery.com>
+	    Cesar Philippidis  <cesar@codesourcery.com>
+
+	* parser.c (cp_parser_omp_clause_name): Add support for finalize
+	and if_present. Make present_or_{copy,copyin,copyout,create} aliases
+	to their non-present_or_* counterparts. Make 'self' an alias to
+	PRAGMA_OACC_CLAUSE_HOST.
+	(cp_parser_oacc_data_clause): Update GOMP mappings for
+	PRAGMA_OACC_CLAUSE_{COPY,COPYIN,COPYOUT,CREATE,DELETE}. Remove
+	PRAGMA_OACC_CLAUSE_{SELF,PRESENT_OR_*}.
+	(cp_parser_oacc_all_clauses): Handle finalize and if_present clauses.
+	Remove support for present_or_* clauses.
+	(OACC_KERNELS_CLAUSE_MASK): Remove PRESENT_OR_* clauses.
+	(OACC_PARALLEL_CLAUSE_MASK): Likewise.
+	(OACC_DECLARE_CLAUSE_MASK): Likewise.
+	(OACC_DATA_CLAUSE_MASK): Likewise.
+	(OACC_ENTER_DATA_CLAUSE_MASK): Remove PRESENT_OR_* clauses.
+	(OACC_EXIT_DATA_CLAUSE_MASK): Add FINALIZE clause.
+	(OACC_UPDATE_CLAUSE_MASK): Remove SELF, add IF_PRESENT.
+	(cp_parser_oacc_declare): Remove PRESENT_OR_* clauses.
+	* pt.c (tsubst_omp_clauses): Handle IF_PRESENT and FINALIZE.
+	* semantics.c (finish_omp_clauses): Handle IF_PRESENT and FINALIZE.
+
 2018-06-20  Marek Polacek  <polacek@redhat.com>
 
 	PR c++/86240
diff --git a/gcc/cp/parser.c b/gcc/cp/parser.c
index b618485..154729c 100644
--- a/gcc/cp/parser.c
+++ b/gcc/cp/parser.c
@@ -31375,6 +31375,8 @@ cp_parser_omp_clause_name (cp_parser *parser)
 	case 'f':
 	  if (!strcmp ("final", p))
 	    result = PRAGMA_OMP_CLAUSE_FINAL;
+	  else if (!strcmp ("finalize", p))
+	    result = PRAGMA_OACC_CLAUSE_FINALIZE;
 	  else if (!strcmp ("firstprivate", p))
 	    result = PRAGMA_OMP_CLAUSE_FIRSTPRIVATE;
 	  else if (!strcmp ("from", p))
@@ -31393,7 +31395,9 @@ cp_parser_omp_clause_name (cp_parser *parser)
 	    result = PRAGMA_OACC_CLAUSE_HOST;
 	  break;
 	case 'i':
-	  if (!strcmp ("inbranch", p))
+	  if (!strcmp ("if_present", p))
+	    result = PRAGMA_OACC_CLAUSE_IF_PRESENT;
+	  else if (!strcmp ("inbranch", p))
 	    result = PRAGMA_OMP_CLAUSE_INBRANCH;
 	  else if (!strcmp ("independent", p))
 	    result = PRAGMA_OACC_CLAUSE_INDEPENDENT;
@@ -31443,16 +31447,16 @@ cp_parser_omp_clause_name (cp_parser *parser)
 	    result = PRAGMA_OACC_CLAUSE_PRESENT;
 	  else if (!strcmp ("present_or_copy", p)
 		   || !strcmp ("pcopy", p))
-	    result = PRAGMA_OACC_CLAUSE_PRESENT_OR_COPY;
+	    result = PRAGMA_OACC_CLAUSE_COPY;
 	  else if (!strcmp ("present_or_copyin", p)
 		   || !strcmp ("pcopyin", p))
-	    result = PRAGMA_OACC_CLAUSE_PRESENT_OR_COPYIN;
+	    result = PRAGMA_OACC_CLAUSE_COPYIN;
 	  else if (!strcmp ("present_or_copyout", p)
 		   || !strcmp ("pcopyout", p))
-	    result = PRAGMA_OACC_CLAUSE_PRESENT_OR_COPYOUT;
+	    result = PRAGMA_OACC_CLAUSE_COPYOUT;
 	  else if (!strcmp ("present_or_create", p)
 		   || !strcmp ("pcreate", p))
-	    result = PRAGMA_OACC_CLAUSE_PRESENT_OR_CREATE;
+	    result = PRAGMA_OACC_CLAUSE_CREATE;
 	  else if (!strcmp ("priority", p))
 	    result = PRAGMA_OMP_CLAUSE_PRIORITY;
 	  else if (!strcmp ("proc_bind", p))
@@ -31469,8 +31473,8 @@ cp_parser_omp_clause_name (cp_parser *parser)
 	    result = PRAGMA_OMP_CLAUSE_SCHEDULE;
 	  else if (!strcmp ("sections", p))
 	    result = PRAGMA_OMP_CLAUSE_SECTIONS;
-	  else if (!strcmp ("self", p))
-	    result = PRAGMA_OACC_CLAUSE_SELF;
+	  else if (!strcmp ("self", p)) /* "self" is a synonym for "host".  */
+	    result = PRAGMA_OACC_CLAUSE_HOST;
 	  else if (!strcmp ("seq", p))
 	    result = PRAGMA_OACC_CLAUSE_SEQ;
 	  else if (!strcmp ("shared", p))
@@ -31730,15 +31734,7 @@ cp_parser_omp_var_list (cp_parser *parser, enum omp_clause_code kind, tree list)
    copyout ( variable-list )
    create ( variable-list )
    delete ( variable-list )
-   present ( variable-list )
-   present_or_copy ( variable-list )
-     pcopy ( variable-list )
-   present_or_copyin ( variable-list )
-     pcopyin ( variable-list )
-   present_or_copyout ( variable-list )
-     pcopyout ( variable-list )
-   present_or_create ( variable-list )
-     pcreate ( variable-list ) */
+   present ( variable-list ) */
 
 static tree
 cp_parser_oacc_data_clause (cp_parser *parser, pragma_omp_clause c_kind,
@@ -31748,19 +31744,19 @@ cp_parser_oacc_data_clause (cp_parser *parser, pragma_omp_clause c_kind,
   switch (c_kind)
     {
     case PRAGMA_OACC_CLAUSE_COPY:
-      kind = GOMP_MAP_FORCE_TOFROM;
+      kind = GOMP_MAP_TOFROM;
       break;
     case PRAGMA_OACC_CLAUSE_COPYIN:
-      kind = GOMP_MAP_FORCE_TO;
+      kind = GOMP_MAP_TO;
       break;
     case PRAGMA_OACC_CLAUSE_COPYOUT:
-      kind = GOMP_MAP_FORCE_FROM;
+      kind = GOMP_MAP_FROM;
       break;
     case PRAGMA_OACC_CLAUSE_CREATE:
-      kind = GOMP_MAP_FORCE_ALLOC;
+      kind = GOMP_MAP_ALLOC;
       break;
     case PRAGMA_OACC_CLAUSE_DELETE:
-      kind = GOMP_MAP_DELETE;
+      kind = GOMP_MAP_RELEASE;
       break;
     case PRAGMA_OACC_CLAUSE_DEVICE:
       kind = GOMP_MAP_FORCE_TO;
@@ -31769,7 +31765,6 @@ cp_parser_oacc_data_clause (cp_parser *parser, pragma_omp_clause c_kind,
       kind = GOMP_MAP_DEVICE_RESIDENT;
       break;
     case PRAGMA_OACC_CLAUSE_HOST:
-    case PRAGMA_OACC_CLAUSE_SELF:
       kind = GOMP_MAP_FORCE_FROM;
       break;
     case PRAGMA_OACC_CLAUSE_LINK:
@@ -31778,18 +31773,6 @@ cp_parser_oacc_data_clause (cp_parser *parser, pragma_omp_clause c_kind,
     case PRAGMA_OACC_CLAUSE_PRESENT:
       kind = GOMP_MAP_FORCE_PRESENT;
       break;
-    case PRAGMA_OACC_CLAUSE_PRESENT_OR_COPY:
-      kind = GOMP_MAP_TOFROM;
-      break;
-    case PRAGMA_OACC_CLAUSE_PRESENT_OR_COPYIN:
-      kind = GOMP_MAP_TO;
-      break;
-    case PRAGMA_OACC_CLAUSE_PRESENT_OR_COPYOUT:
-      kind = GOMP_MAP_FROM;
-      break;
-    case PRAGMA_OACC_CLAUSE_PRESENT_OR_CREATE:
-      kind = GOMP_MAP_ALLOC;
-      break;
     default:
       gcc_unreachable ();
     }
@@ -31828,8 +31811,9 @@ cp_parser_oacc_data_clause_deviceptr (cp_parser *parser, tree list)
   return list;
 }
 
-/* OpenACC 2.0:
+/* OpenACC 2.5:
    auto
+   finalize
    independent
    nohost
    seq */
@@ -33794,6 +33778,11 @@ cp_parser_oacc_all_clauses (cp_parser *parser, omp_clause_mask mask,
 	  clauses = cp_parser_oacc_data_clause (parser, c_kind, clauses);
 	  c_name = "device_resident";
 	  break;
+	case PRAGMA_OACC_CLAUSE_FINALIZE:
+	  clauses = cp_parser_oacc_simple_clause (parser, OMP_CLAUSE_FINALIZE,
+						  clauses, here);
+	  c_name = "finalize";
+	  break;
 	case PRAGMA_OACC_CLAUSE_FIRSTPRIVATE:
 	  clauses = cp_parser_omp_var_list (parser, OMP_CLAUSE_FIRSTPRIVATE,
 					    clauses);
@@ -33812,6 +33801,12 @@ cp_parser_oacc_all_clauses (cp_parser *parser, omp_clause_mask mask,
 	  clauses = cp_parser_omp_clause_if (parser, clauses, here, false);
 	  c_name = "if";
 	  break;
+	case PRAGMA_OACC_CLAUSE_IF_PRESENT:
+	  clauses = cp_parser_oacc_simple_clause (parser,
+						  OMP_CLAUSE_IF_PRESENT,
+						  clauses, here);
+	  c_name = "if_present";
+	  break;
 	case PRAGMA_OACC_CLAUSE_INDEPENDENT:
 	  clauses = cp_parser_oacc_simple_clause (parser,
 						  OMP_CLAUSE_INDEPENDENT,
@@ -33838,22 +33833,6 @@ cp_parser_oacc_all_clauses (cp_parser *parser, omp_clause_mask mask,
 	  clauses = cp_parser_oacc_data_clause (parser, c_kind, clauses);
 	  c_name = "present";
 	  break;
-	case PRAGMA_OACC_CLAUSE_PRESENT_OR_COPY:
-	  clauses = cp_parser_oacc_data_clause (parser, c_kind, clauses);
-	  c_name = "present_or_copy";
-	  break;
-	case PRAGMA_OACC_CLAUSE_PRESENT_OR_COPYIN:
-	  clauses = cp_parser_oacc_data_clause (parser, c_kind, clauses);
-	  c_name = "present_or_copyin";
-	  break;
-	case PRAGMA_OACC_CLAUSE_PRESENT_OR_COPYOUT:
-	  clauses = cp_parser_oacc_data_clause (parser, c_kind, clauses);
-	  c_name = "present_or_copyout";
-	  break;
-	case PRAGMA_OACC_CLAUSE_PRESENT_OR_CREATE:
-	  clauses = cp_parser_oacc_data_clause (parser, c_kind, clauses);
-	  c_name = "present_or_create";
-	  break;
 	case PRAGMA_OACC_CLAUSE_PRIVATE:
 	  clauses = cp_parser_omp_var_list (parser, OMP_CLAUSE_PRIVATE,
 					    clauses);
@@ -33863,10 +33842,6 @@ cp_parser_oacc_all_clauses (cp_parser *parser, omp_clause_mask mask,
 	  clauses = cp_parser_omp_clause_reduction (parser, clauses);
 	  c_name = "reduction";
 	  break;
-	case PRAGMA_OACC_CLAUSE_SELF:
-	  clauses = cp_parser_oacc_data_clause (parser, c_kind, clauses);
-	  c_name = "self";
-	  break;
 	case PRAGMA_OACC_CLAUSE_SEQ:
 	  clauses = cp_parser_oacc_simple_clause (parser, OMP_CLAUSE_SEQ,
 						 clauses, here);
@@ -36802,11 +36777,7 @@ cp_parser_oacc_cache (cp_parser *parser, cp_token *pragma_tok)
 	| (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_PRESENT) )
 
 static tree
 cp_parser_oacc_data (cp_parser *parser, cp_token *pragma_tok, bool *if_p)
@@ -36861,11 +36832,7 @@ cp_parser_oacc_host_data (cp_parser *parser, cp_token *pragma_tok, bool *if_p)
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_DEVICEPTR)		\
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_DEVICE_RESIDENT)	\
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_LINK)		\
-	| (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_PRESENT) )
 
 static tree
 cp_parser_oacc_declare (cp_parser *parser, cp_token *pragma_tok)
@@ -36898,8 +36865,8 @@ cp_parser_oacc_declare (cp_parser *parser, cp_token *pragma_tok)
       switch (OMP_CLAUSE_MAP_KIND (t))
 	{
 	case GOMP_MAP_FIRSTPRIVATE_POINTER:
-	case GOMP_MAP_FORCE_ALLOC:
-	case GOMP_MAP_FORCE_TO:
+	case GOMP_MAP_ALLOC:
+	case GOMP_MAP_TO:
 	case GOMP_MAP_FORCE_DEVICEPTR:
 	case GOMP_MAP_DEVICE_RESIDENT:
 	  break;
@@ -37010,8 +36977,6 @@ cp_parser_oacc_declare (cp_parser *parser, cp_token *pragma_tok)
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_ASYNC)		\
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_COPYIN)		\
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_CREATE)		\
-	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_PRESENT_OR_COPYIN)	\
-	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_PRESENT_OR_CREATE)	\
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_WAIT) )
 
 #define OACC_EXIT_DATA_CLAUSE_MASK					\
@@ -37019,6 +36984,7 @@ cp_parser_oacc_declare (cp_parser *parser, cp_token *pragma_tok)
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_ASYNC)		\
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_COPYOUT)		\
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_DELETE) 		\
+	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_FINALIZE) 		\
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_WAIT) )
 
 static tree
@@ -37131,10 +37097,6 @@ cp_parser_oacc_loop (cp_parser *parser, cp_token *pragma_tok, char *p_name,
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_NUM_GANGS)		\
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_NUM_WORKERS)		\
 	| (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_VECTOR_LENGTH)	\
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_WAIT) )
 
@@ -37151,10 +37113,6 @@ cp_parser_oacc_loop (cp_parser *parser, cp_token *pragma_tok, char *p_name,
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_NUM_GANGS)		\
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_NUM_WORKERS)		\
 	| (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_PRIVATE)		\
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_REDUCTION)		\
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_VECTOR_LENGTH)       \
@@ -37215,7 +37173,7 @@ cp_parser_oacc_kernels_parallel (cp_parser *parser, cp_token *pragma_tok,
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_DEVICE)		\
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_HOST)		\
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_IF)			\
-	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_SELF)		\
+	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_IF_PRESENT)		\
 	| (OMP_CLAUSE_MASK_1 << PRAGMA_OACC_CLAUSE_WAIT))
 
 static tree
diff --git a/gcc/cp/pt.c b/gcc/cp/pt.c
index be42b20..c5433dc 100644
--- a/gcc/cp/pt.c
+++ b/gcc/cp/pt.c
@@ -16107,6 +16107,8 @@ tsubst_omp_clauses (tree clauses, enum c_omp_region_type ort,
 	case OMP_CLAUSE_INDEPENDENT:
 	case OMP_CLAUSE_AUTO:
 	case OMP_CLAUSE_SEQ:
+	case OMP_CLAUSE_IF_PRESENT:
+	case OMP_CLAUSE_FINALIZE:
 	  break;
 	default:
 	  gcc_unreachable ();
diff --git a/gcc/cp/semantics.c b/gcc/cp/semantics.c
index 38b7b66..bad712e 100644
--- a/gcc/cp/semantics.c
+++ b/gcc/cp/semantics.c
@@ -7091,6 +7091,8 @@ finish_omp_clauses (tree clauses, enum c_omp_region_type ort)
 	case OMP_CLAUSE_AUTO:
 	case OMP_CLAUSE_INDEPENDENT:
 	case OMP_CLAUSE_SEQ:
+	case OMP_CLAUSE_IF_PRESENT:
+	case OMP_CLAUSE_FINALIZE:
 	  break;
 
 	case OMP_CLAUSE_TILE:
diff --git a/gcc/fortran/ChangeLog b/gcc/fortran/ChangeLog
index 0ea7a51..99a3114 100644
--- a/gcc/fortran/ChangeLog
+++ b/gcc/fortran/ChangeLog
@@ -1,3 +1,25 @@
+2018-06-20  Chung-Lin Tang <cltang@codesourcery.com>
+	    Thomas Schwinge <thomas@codesourcery.com>
+	    Cesar Philippidis  <cesar@codesourcery.com>
+
+	* gfortran.h (gfc_omp_clauses): Add unsigned if_present, finalize
+	bitfields.
+	* openmp.c (enum omp_mask2): Remove OMP_CLAUSE_PRESENT_OR_*. Add
+	OMP_CLAUSE_{IF_PRESENT,FINALIZE}.
+	(gfc_match_omp_clauses): Update handling of copy, copyin, copyout,
+	create, deviceptr, present_of_*. Add support for finalize and
+	if_present.
+	(OACC_PARALLEL_CLAUSES): Remove PRESENT_OR_* clauses.
+	(OACC_KERNELS_CLAUSES): Likewise.
+	(OACC_DATA_CLAUSES): Likewise.
+	(OACC_DECLARE_CLAUSES): Likewise.
+	(OACC_UPDATE_CLAUSES): Add IF_PRESENT clause.
+	(OACC_ENTER_DATA_CLAUSES): Remove PRESENT_OR_* clauses.
+	(OACC_EXIT_DATA_CLAUSES): Add FINALIZE clause.
+	(gfc_match_oacc_declare): Update to OpenACC 2.5 semantics.
+	* trans-openmp.c (gfc_trans_omp_clauses): Add support for IF_PRESENT
+	and FINALIZE.
+
 2018-06-18  Eric Botcazou  <ebotcazou@adacore.com>
 
 	* trans-decl.c (gfc_get_fake_result_decl): Revert latest change.
diff --git a/gcc/fortran/gfortran.h b/gcc/fortran/gfortran.h
index 1d98d25..0b89f8d 100644
--- a/gcc/fortran/gfortran.h
+++ b/gcc/fortran/gfortran.h
@@ -1344,6 +1344,7 @@ typedef struct gfc_omp_clauses
   gfc_expr_list *tile_list;
   unsigned async:1, gang:1, worker:1, vector:1, seq:1, independent:1;
   unsigned wait:1, par_auto:1, gang_static:1;
+  unsigned if_present:1, finalize:1;
   locus loc;
 
 }
diff --git a/gcc/fortran/openmp.c b/gcc/fortran/openmp.c
index 97d6e78..94a7f7e 100644
--- a/gcc/fortran/openmp.c
+++ b/gcc/fortran/openmp.c
@@ -796,10 +796,6 @@ enum omp_mask2
   OMP_CLAUSE_COPYOUT,
   OMP_CLAUSE_CREATE,
   OMP_CLAUSE_PRESENT,
-  OMP_CLAUSE_PRESENT_OR_COPY,
-  OMP_CLAUSE_PRESENT_OR_COPYIN,
-  OMP_CLAUSE_PRESENT_OR_COPYOUT,
-  OMP_CLAUSE_PRESENT_OR_CREATE,
   OMP_CLAUSE_DEVICEPTR,
   OMP_CLAUSE_GANG,
   OMP_CLAUSE_WORKER,
@@ -813,6 +809,8 @@ enum omp_mask2
   OMP_CLAUSE_DELETE,
   OMP_CLAUSE_AUTO,
   OMP_CLAUSE_TILE,
+  OMP_CLAUSE_IF_PRESENT,
+  OMP_CLAUSE_FINALIZE,
   /* This must come last.  */
   OMP_MASK2_LAST
 };
@@ -1041,7 +1039,7 @@ gfc_match_omp_clauses (gfc_omp_clauses **cp, const omp_mask mask,
 	  if ((mask & OMP_CLAUSE_COPY)
 	      && gfc_match ("copy ( ") == MATCH_YES
 	      && gfc_match_omp_map_clause (&c->lists[OMP_LIST_MAP],
-					   OMP_MAP_FORCE_TOFROM))
+					   OMP_MAP_TOFROM))
 	    continue;
 	  if (mask & OMP_CLAUSE_COPYIN)
 	    {
@@ -1049,7 +1047,7 @@ gfc_match_omp_clauses (gfc_omp_clauses **cp, const omp_mask mask,
 		{
 		  if (gfc_match ("copyin ( ") == MATCH_YES
 		      && gfc_match_omp_map_clause (&c->lists[OMP_LIST_MAP],
-						   OMP_MAP_FORCE_TO))
+						   OMP_MAP_TO))
 		    continue;
 		}
 	      else if (gfc_match_omp_variable_list ("copyin (",
@@ -1060,7 +1058,7 @@ gfc_match_omp_clauses (gfc_omp_clauses **cp, const omp_mask mask,
 	  if ((mask & OMP_CLAUSE_COPYOUT)
 	      && gfc_match ("copyout ( ") == MATCH_YES
 	      && gfc_match_omp_map_clause (&c->lists[OMP_LIST_MAP],
-					   OMP_MAP_FORCE_FROM))
+					   OMP_MAP_FROM))
 	    continue;
 	  if ((mask & OMP_CLAUSE_COPYPRIVATE)
 	      && gfc_match_omp_variable_list ("copyprivate (",
@@ -1070,7 +1068,7 @@ gfc_match_omp_clauses (gfc_omp_clauses **cp, const omp_mask mask,
 	  if ((mask & OMP_CLAUSE_CREATE)
 	      && gfc_match ("create ( ") == MATCH_YES
 	      && gfc_match_omp_map_clause (&c->lists[OMP_LIST_MAP],
-					   OMP_MAP_FORCE_ALLOC))
+					   OMP_MAP_ALLOC))
 	    continue;
 	  break;
 	case 'd':
@@ -1106,7 +1104,7 @@ gfc_match_omp_clauses (gfc_omp_clauses **cp, const omp_mask mask,
 	  if ((mask & OMP_CLAUSE_DELETE)
 	      && gfc_match ("delete ( ") == MATCH_YES
 	      && gfc_match_omp_map_clause (&c->lists[OMP_LIST_MAP],
-					   OMP_MAP_DELETE))
+					   OMP_MAP_RELEASE))
 	    continue;
 	  if ((mask & OMP_CLAUSE_DEPEND)
 	      && gfc_match ("depend ( ") == MATCH_YES)
@@ -1161,19 +1159,10 @@ gfc_match_omp_clauses (gfc_omp_clauses **cp, const omp_mask mask,
 					   OMP_MAP_FORCE_TO))
 	    continue;
 	  if ((mask & OMP_CLAUSE_DEVICEPTR)
-	      && gfc_match ("deviceptr ( ") == MATCH_YES)
-	    {
-	      gfc_omp_namelist **list = &c->lists[OMP_LIST_MAP];
-	      gfc_omp_namelist **head = NULL;
-	      if (gfc_match_omp_variable_list ("", list, true, NULL,
-					       &head, false) == MATCH_YES)
-		{
-		  gfc_omp_namelist *n;
-		  for (n = *head; n; n = n->next)
-		    n->u.map_op = OMP_MAP_FORCE_DEVICEPTR;
-		  continue;
-		}
-	    }
+	      && gfc_match ("deviceptr ( ") == MATCH_YES
+	      && gfc_match_omp_map_clause (&c->lists[OMP_LIST_MAP],
+					   OMP_MAP_FORCE_DEVICEPTR))
+	    continue;
 	  if ((mask & OMP_CLAUSE_DEVICE_RESIDENT)
 	      && gfc_match_omp_variable_list
 		   ("device_resident (",
@@ -1202,6 +1191,14 @@ gfc_match_omp_clauses (gfc_omp_clauses **cp, const omp_mask mask,
 	      && c->final_expr == NULL
 	      && gfc_match ("final ( %e )", &c->final_expr) == MATCH_YES)
 	    continue;
+	  if ((mask & OMP_CLAUSE_FINALIZE)
+	      && !c->finalize
+	      && gfc_match ("finalize") == MATCH_YES)
+	    {
+	      c->finalize = true;
+	      needs_space = true;
+	      continue;
+	    }
 	  if ((mask & OMP_CLAUSE_FIRSTPRIVATE)
 	      && gfc_match_omp_variable_list ("firstprivate (",
 					      &c->lists[OMP_LIST_FIRSTPRIVATE],
@@ -1274,6 +1271,14 @@ gfc_match_omp_clauses (gfc_omp_clauses **cp, const omp_mask mask,
 		}
 	      gfc_current_locus = old_loc;
 	    }
+	  if ((mask & OMP_CLAUSE_IF_PRESENT)
+	      && !c->if_present
+	      && gfc_match ("if_present") == MATCH_YES)
+	    {
+	      c->if_present = true;
+	      needs_space = true;
+	      continue;
+	    }
 	  if ((mask & OMP_CLAUSE_INBRANCH)
 	      && !c->inbranch
 	      && !c->notinbranch
@@ -1503,22 +1508,22 @@ gfc_match_omp_clauses (gfc_omp_clauses **cp, const omp_mask mask,
 	    }
 	  break;
 	case 'p':
-	  if ((mask & OMP_CLAUSE_PRESENT_OR_COPY)
+	  if ((mask & OMP_CLAUSE_COPY)
 	      && gfc_match ("pcopy ( ") == MATCH_YES
 	      && gfc_match_omp_map_clause (&c->lists[OMP_LIST_MAP],
 					   OMP_MAP_TOFROM))
 	    continue;
-	  if ((mask & OMP_CLAUSE_PRESENT_OR_COPYIN)
+	  if ((mask & OMP_CLAUSE_COPYIN)
 	      && gfc_match ("pcopyin ( ") == MATCH_YES
 	      && gfc_match_omp_map_clause (&c->lists[OMP_LIST_MAP],
 					   OMP_MAP_TO))
 	    continue;
-	  if ((mask & OMP_CLAUSE_PRESENT_OR_COPYOUT)
+	  if ((mask & OMP_CLAUSE_COPYOUT)
 	      && gfc_match ("pcopyout ( ") == MATCH_YES
 	      && gfc_match_omp_map_clause (&c->lists[OMP_LIST_MAP],
 					   OMP_MAP_FROM))
 	    continue;
-	  if ((mask & OMP_CLAUSE_PRESENT_OR_CREATE)
+	  if ((mask & OMP_CLAUSE_CREATE)
 	      && gfc_match ("pcreate ( ") == MATCH_YES
 	      && gfc_match_omp_map_clause (&c->lists[OMP_LIST_MAP],
 					   OMP_MAP_ALLOC))
@@ -1528,22 +1533,22 @@ gfc_match_omp_clauses (gfc_omp_clauses **cp, const omp_mask mask,
 	      && gfc_match_omp_map_clause (&c->lists[OMP_LIST_MAP],
 					   OMP_MAP_FORCE_PRESENT))
 	    continue;
-	  if ((mask & OMP_CLAUSE_PRESENT_OR_COPY)
+	  if ((mask & OMP_CLAUSE_COPY)
 	      && gfc_match ("present_or_copy ( ") == MATCH_YES
 	      && gfc_match_omp_map_clause (&c->lists[OMP_LIST_MAP],
 					   OMP_MAP_TOFROM))
 	    continue;
-	  if ((mask & OMP_CLAUSE_PRESENT_OR_COPYIN)
+	  if ((mask & OMP_CLAUSE_COPYIN)
 	      && gfc_match ("present_or_copyin ( ") == MATCH_YES
 	      && gfc_match_omp_map_clause (&c->lists[OMP_LIST_MAP],
 					   OMP_MAP_TO))
 	    continue;
-	  if ((mask & OMP_CLAUSE_PRESENT_OR_COPYOUT)
+	  if ((mask & OMP_CLAUSE_COPYOUT)
 	      && gfc_match ("present_or_copyout ( ") == MATCH_YES
 	      && gfc_match_omp_map_clause (&c->lists[OMP_LIST_MAP],
 					   OMP_MAP_FROM))
 	    continue;
-	  if ((mask & OMP_CLAUSE_PRESENT_OR_CREATE)
+	  if ((mask & OMP_CLAUSE_CREATE)
 	      && gfc_match ("present_or_create ( ") == MATCH_YES
 	      && gfc_match_omp_map_clause (&c->lists[OMP_LIST_MAP],
 					   OMP_MAP_ALLOC))
@@ -1925,23 +1930,19 @@ gfc_match_omp_clauses (gfc_omp_clauses **cp, const omp_mask mask,
   (omp_mask (OMP_CLAUSE_IF) | OMP_CLAUSE_ASYNC | OMP_CLAUSE_NUM_GANGS	      \
    | OMP_CLAUSE_NUM_WORKERS | OMP_CLAUSE_VECTOR_LENGTH | OMP_CLAUSE_REDUCTION \
    | OMP_CLAUSE_COPY | OMP_CLAUSE_COPYIN | OMP_CLAUSE_COPYOUT		      \
-   | OMP_CLAUSE_CREATE | OMP_CLAUSE_PRESENT | OMP_CLAUSE_PRESENT_OR_COPY      \
-   | OMP_CLAUSE_PRESENT_OR_COPYIN | OMP_CLAUSE_PRESENT_OR_COPYOUT	      \
-   | OMP_CLAUSE_PRESENT_OR_CREATE | OMP_CLAUSE_DEVICEPTR | OMP_CLAUSE_PRIVATE \
-   | OMP_CLAUSE_FIRSTPRIVATE | OMP_CLAUSE_DEFAULT | OMP_CLAUSE_WAIT)
+   | OMP_CLAUSE_CREATE | OMP_CLAUSE_PRESENT | OMP_CLAUSE_DEVICEPTR	      \
+   | OMP_CLAUSE_PRIVATE | OMP_CLAUSE_FIRSTPRIVATE | OMP_CLAUSE_DEFAULT	      \
+   | OMP_CLAUSE_WAIT)
 #define OACC_KERNELS_CLAUSES \
   (omp_mask (OMP_CLAUSE_IF) | OMP_CLAUSE_ASYNC | OMP_CLAUSE_NUM_GANGS	      \
    | OMP_CLAUSE_NUM_WORKERS | OMP_CLAUSE_VECTOR_LENGTH | OMP_CLAUSE_DEVICEPTR \
    | OMP_CLAUSE_COPY | OMP_CLAUSE_COPYIN | OMP_CLAUSE_COPYOUT		      \
-   | OMP_CLAUSE_CREATE | OMP_CLAUSE_PRESENT | OMP_CLAUSE_PRESENT_OR_COPY      \
-   | OMP_CLAUSE_PRESENT_OR_COPYIN | OMP_CLAUSE_PRESENT_OR_COPYOUT	      \
-   | OMP_CLAUSE_PRESENT_OR_CREATE | OMP_CLAUSE_DEFAULT | OMP_CLAUSE_WAIT)
+   | OMP_CLAUSE_CREATE | OMP_CLAUSE_PRESENT | OMP_CLAUSE_DEFAULT	      \
+   | OMP_CLAUSE_WAIT)
 #define OACC_DATA_CLAUSES \
   (omp_mask (OMP_CLAUSE_IF) | OMP_CLAUSE_DEVICEPTR  | OMP_CLAUSE_COPY	      \
    | OMP_CLAUSE_COPYIN | OMP_CLAUSE_COPYOUT | OMP_CLAUSE_CREATE		      \
-   | OMP_CLAUSE_PRESENT | OMP_CLAUSE_PRESENT_OR_COPY			      \
-   | OMP_CLAUSE_PRESENT_OR_COPYIN | OMP_CLAUSE_PRESENT_OR_COPYOUT	      \
-   | OMP_CLAUSE_PRESENT_OR_CREATE)
+   | OMP_CLAUSE_PRESENT)
 #define OACC_LOOP_CLAUSES \
   (omp_mask (OMP_CLAUSE_COLLAPSE) | OMP_CLAUSE_GANG | OMP_CLAUSE_WORKER	      \
    | OMP_CLAUSE_VECTOR | OMP_CLAUSE_SEQ | OMP_CLAUSE_INDEPENDENT	      \
@@ -1955,19 +1956,17 @@ gfc_match_omp_clauses (gfc_omp_clauses **cp, const omp_mask mask,
 #define OACC_DECLARE_CLAUSES \
   (omp_mask (OMP_CLAUSE_COPY) | OMP_CLAUSE_COPYIN | OMP_CLAUSE_COPYOUT	      \
    | OMP_CLAUSE_CREATE | OMP_CLAUSE_DEVICEPTR | OMP_CLAUSE_DEVICE_RESIDENT    \
-   | OMP_CLAUSE_PRESENT | OMP_CLAUSE_PRESENT_OR_COPY			      \
-   | OMP_CLAUSE_PRESENT_OR_COPYIN | OMP_CLAUSE_PRESENT_OR_COPYOUT	      \
-   | OMP_CLAUSE_PRESENT_OR_CREATE | OMP_CLAUSE_LINK)
+   | OMP_CLAUSE_PRESENT			      \
+   | OMP_CLAUSE_LINK)
 #define OACC_UPDATE_CLAUSES \
   (omp_mask (OMP_CLAUSE_IF) | OMP_CLAUSE_ASYNC | OMP_CLAUSE_HOST_SELF	      \
-   | OMP_CLAUSE_DEVICE | OMP_CLAUSE_WAIT)
+   | OMP_CLAUSE_DEVICE | OMP_CLAUSE_WAIT | OMP_CLAUSE_IF_PRESENT)
 #define OACC_ENTER_DATA_CLAUSES \
   (omp_mask (OMP_CLAUSE_IF) | OMP_CLAUSE_ASYNC | OMP_CLAUSE_WAIT	      \
-   | OMP_CLAUSE_COPYIN | OMP_CLAUSE_CREATE | OMP_CLAUSE_PRESENT_OR_COPYIN     \
-   | OMP_CLAUSE_PRESENT_OR_CREATE)
+   | OMP_CLAUSE_COPYIN | OMP_CLAUSE_CREATE)
 #define OACC_EXIT_DATA_CLAUSES \
   (omp_mask (OMP_CLAUSE_IF) | OMP_CLAUSE_ASYNC | OMP_CLAUSE_WAIT	      \
-   | OMP_CLAUSE_COPYOUT | OMP_CLAUSE_DELETE)
+   | OMP_CLAUSE_COPYOUT | OMP_CLAUSE_DELETE | OMP_CLAUSE_FINALIZE)
 #define OACC_WAIT_CLAUSES \
   omp_mask (OMP_CLAUSE_ASYNC)
 #define OACC_ROUTINE_CLAUSES \
@@ -2061,8 +2060,7 @@ gfc_match_oacc_declare (void)
 
       if (s->ns->proc_name && s->ns->proc_name->attr.proc == PROC_MODULE)
 	{
-	  if (n->u.map_op != OMP_MAP_FORCE_ALLOC
-	      && n->u.map_op != OMP_MAP_FORCE_TO)
+	  if (n->u.map_op != OMP_MAP_ALLOC && n->u.map_op != OMP_MAP_TO)
 	    {
 	      gfc_error ("Invalid clause in module with !$ACC DECLARE at %L",
 			 &where);
@@ -2072,6 +2070,13 @@ gfc_match_oacc_declare (void)
 	  module_var = true;
 	}
 
+      if (ns->proc_name->attr.oacc_function)
+	{
+	  gfc_error ("Invalid declare in routine with $!ACC DECLARE at %L",
+		     &where);
+	  return MATCH_ERROR;
+	}
+
       if (s->attr.use_assoc)
 	{
 	  gfc_error ("Variable is USE-associated with !$ACC DECLARE at %L",
@@ -2090,10 +2095,12 @@ gfc_match_oacc_declare (void)
       switch (n->u.map_op)
 	{
 	  case OMP_MAP_FORCE_ALLOC:
+	  case OMP_MAP_ALLOC:
 	    s->attr.oacc_declare_create = 1;
 	    break;
 
 	  case OMP_MAP_FORCE_TO:
+	  case OMP_MAP_TO:
 	    s->attr.oacc_declare_copyin = 1;
 	    break;
 
diff --git a/gcc/fortran/trans-openmp.c b/gcc/fortran/trans-openmp.c
index 795175d..f038f4c 100644
--- a/gcc/fortran/trans-openmp.c
+++ b/gcc/fortran/trans-openmp.c
@@ -2895,6 +2895,16 @@ gfc_trans_omp_clauses (stmtblock_t *block, gfc_omp_clauses *clauses,
       c = build_omp_clause (where.lb->location, OMP_CLAUSE_AUTO);
       omp_clauses = gfc_trans_add_clause (c, omp_clauses);
     }
+  if (clauses->if_present)
+    {
+      c = build_omp_clause (where.lb->location, OMP_CLAUSE_IF_PRESENT);
+      omp_clauses = gfc_trans_add_clause (c, omp_clauses);
+    }
+  if (clauses->finalize)
+    {
+      c = build_omp_clause (where.lb->location, OMP_CLAUSE_FINALIZE);
+      omp_clauses = gfc_trans_add_clause (c, omp_clauses);
+    }
   if (clauses->independent)
     {
       c = build_omp_clause (where.lb->location, OMP_CLAUSE_INDEPENDENT);
diff --git a/gcc/gimplify.c b/gcc/gimplify.c
index 1523a27..97543ed 100644
--- a/gcc/gimplify.c
+++ b/gcc/gimplify.c
@@ -8524,6 +8524,8 @@ gimplify_scan_omp_clauses (tree *list_p, gimple_seq *pre_p,
 	case OMP_CLAUSE_NOGROUP:
 	case OMP_CLAUSE_THREADS:
 	case OMP_CLAUSE_SIMD:
+	case OMP_CLAUSE_IF_PRESENT:
+	case OMP_CLAUSE_FINALIZE:
 	  break;
 
 	case OMP_CLAUSE_DEFAULTMAP:
@@ -9305,6 +9307,8 @@ gimplify_adjust_omp_clauses (gimple_seq *pre_p, gimple_seq body, tree *list_p,
 	case OMP_CLAUSE_AUTO:
 	case OMP_CLAUSE_SEQ:
 	case OMP_CLAUSE_TILE:
+	case OMP_CLAUSE_IF_PRESENT:
+	case OMP_CLAUSE_FINALIZE:
 	  break;
 
 	default:
@@ -9361,21 +9365,7 @@ gimplify_oacc_declare_1 (tree clause)
   switch (kind)
     {
       case GOMP_MAP_ALLOC:
-      case GOMP_MAP_FORCE_ALLOC:
-      case GOMP_MAP_FORCE_TO:
-	new_op = GOMP_MAP_DELETE;
-	ret = true;
-	break;
-
-      case GOMP_MAP_FORCE_FROM:
-	OMP_CLAUSE_SET_MAP_KIND (clause, GOMP_MAP_FORCE_ALLOC);
-	new_op = GOMP_MAP_FORCE_FROM;
-	ret = true;
-	break;
-
-      case GOMP_MAP_FORCE_TOFROM:
-	OMP_CLAUSE_SET_MAP_KIND (clause, GOMP_MAP_FORCE_TO);
-	new_op = GOMP_MAP_FORCE_FROM;
+	new_op = GOMP_MAP_RELEASE;
 	ret = true;
 	break;
 
@@ -10817,6 +10807,53 @@ gimplify_omp_target_update (tree *expr_p, gimple_seq *pre_p)
 			     ort, TREE_CODE (expr));
   gimplify_adjust_omp_clauses (pre_p, NULL, &OMP_STANDALONE_CLAUSES (expr),
 			       TREE_CODE (expr));
+  if (TREE_CODE (expr) == OACC_UPDATE
+      && omp_find_clause (OMP_STANDALONE_CLAUSES (expr),
+			  OMP_CLAUSE_IF_PRESENT))
+    {
+      /* The runtime uses GOMP_MAP_{TO,FROM} to denote the if_present
+	 clause.  */
+      for (tree c = OMP_STANDALONE_CLAUSES (expr); c; c = OMP_CLAUSE_CHAIN (c))
+	if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_MAP)
+	  switch (OMP_CLAUSE_MAP_KIND (c))
+	    {
+	    case GOMP_MAP_FORCE_TO:
+	      OMP_CLAUSE_SET_MAP_KIND (c, GOMP_MAP_TO);
+	      break;
+	    case GOMP_MAP_FORCE_FROM:
+	      OMP_CLAUSE_SET_MAP_KIND (c, GOMP_MAP_FROM);
+	      break;
+	    default:
+	      break;
+	    }
+    }
+  else if (TREE_CODE (expr) == OACC_EXIT_DATA
+	   && omp_find_clause (OMP_STANDALONE_CLAUSES (expr),
+			       OMP_CLAUSE_FINALIZE))
+    {
+      /* Use GOMP_MAP_DELETE/GOMP_MAP_FORCE_FROM to denote that "finalize"
+	 semantics apply to all mappings of this OpenACC directive.  */
+      bool finalize_marked = false;
+      for (tree c = OMP_STANDALONE_CLAUSES (expr); c; c = OMP_CLAUSE_CHAIN (c))
+	if (OMP_CLAUSE_CODE (c) == OMP_CLAUSE_MAP)
+	  switch (OMP_CLAUSE_MAP_KIND (c))
+	    {
+	    case GOMP_MAP_FROM:
+	      OMP_CLAUSE_SET_MAP_KIND (c, GOMP_MAP_FORCE_FROM);
+	      finalize_marked = true;
+	      break;
+	    case GOMP_MAP_RELEASE:
+	      OMP_CLAUSE_SET_MAP_KIND (c, GOMP_MAP_DELETE);
+	      finalize_marked = true;
+	      break;
+	    default:
+	      /* Check consistency: libgomp relies on the very first data
+		 mapping clause being marked, so make sure we did that before
+		 any other mapping clauses.  */
+	      gcc_assert (finalize_marked);
+	      break;
+	    }
+    }
   stmt = gimple_build_omp_target (NULL, kind, OMP_STANDALONE_CLAUSES (expr));
 
   gimplify_seq_add_stmt (pre_p, stmt);
diff --git a/gcc/omp-low.c b/gcc/omp-low.c
index ba6c705..c591231 100644
--- a/gcc/omp-low.c
+++ b/gcc/omp-low.c
@@ -642,8 +642,7 @@ build_sender_ref (tree var, omp_context *ctx)
    BASE_POINTERS_RESTRICT, declare the field with restrict.  */
 
 static void
-install_var_field (tree var, bool by_ref, int mask, omp_context *ctx,
-		   bool base_pointers_restrict = false)
+install_var_field (tree var, bool by_ref, int mask, omp_context *ctx)
 {
   tree field, type, sfield = NULL_TREE;
   splay_tree_key key = (splay_tree_key) var;
@@ -674,11 +673,7 @@ install_var_field (tree var, bool by_ref, int mask, omp_context *ctx,
       type = build_pointer_type (build_pointer_type (type));
     }
   else if (by_ref)
-    {
-      type = build_pointer_type (type);
-      if (base_pointers_restrict)
-	type = build_qualified_type (type, TYPE_QUAL_RESTRICT);
-    }
+    type = build_pointer_type (type);
   else if ((mask & 3) == 1 && omp_is_reference (var))
     type = TREE_TYPE (type);
 
@@ -992,12 +987,10 @@ fixup_child_record_type (omp_context *ctx)
 }
 
 /* Instantiate decls as necessary in CTX to satisfy the data sharing
-   specified by CLAUSES.  If BASE_POINTERS_RESTRICT, install var field with
-   restrict.  */
+   specified by CLAUSES.  */
 
 static void
-scan_sharing_clauses (tree clauses, omp_context *ctx,
-		      bool base_pointers_restrict = false)
+scan_sharing_clauses (tree clauses, omp_context *ctx)
 {
   tree c, decl;
   bool scan_array_reductions = false;
@@ -1256,8 +1249,7 @@ scan_sharing_clauses (tree clauses, omp_context *ctx,
 		      && TREE_CODE (TREE_TYPE (decl)) == ARRAY_TYPE)
 		    install_var_field (decl, true, 7, ctx);
 		  else
-		    install_var_field (decl, true, 3, ctx,
-				       base_pointers_restrict);
+		    install_var_field (decl, true, 3, ctx);
 		  if (is_gimple_omp_offloaded (ctx->stmt)
 		      && !OMP_CLAUSE_MAP_IN_REDUCTION (c))
 		    install_var_local (decl, ctx);
@@ -1328,6 +1320,8 @@ scan_sharing_clauses (tree clauses, omp_context *ctx,
 	case OMP_CLAUSE_TILE:
 	case OMP_CLAUSE__SIMT_:
 	case OMP_CLAUSE_DEFAULT:
+	case OMP_CLAUSE_IF_PRESENT:
+	case OMP_CLAUSE_FINALIZE:
 	  break;
 
 	case OMP_CLAUSE_ALIGNED:
@@ -1499,6 +1493,8 @@ scan_sharing_clauses (tree clauses, omp_context *ctx,
 	case OMP_CLAUSE_TILE:
 	case OMP_CLAUSE__GRIDDIM_:
 	case OMP_CLAUSE__SIMT_:
+	case OMP_CLAUSE_IF_PRESENT:
+	case OMP_CLAUSE_FINALIZE:
 	  break;
 
 	case OMP_CLAUSE__CACHE_:
@@ -2266,68 +2262,6 @@ scan_omp_single (gomp_single *stmt, omp_context *outer_ctx)
     layout_type (ctx->record_type);
 }
 
-/* Return true if the CLAUSES of an omp target guarantee that the base pointers
-   used in the corresponding offloaded function are restrict.  */
-
-static bool
-omp_target_base_pointers_restrict_p (tree clauses)
-{
-  /* The analysis relies on the GOMP_MAP_FORCE_* mapping kinds, which are only
-     used by OpenACC.  */
-  if (flag_openacc == 0)
-    return false;
-
-  /* I.  Basic example:
-
-       void foo (void)
-       {
-	 unsigned int a[2], b[2];
-
-	 #pragma acc kernels \
-	   copyout (a) \
-	   copyout (b)
-	 {
-	   a[0] = 0;
-	   b[0] = 1;
-	 }
-       }
-
-     After gimplification, we have:
-
-       #pragma omp target oacc_kernels \
-	 map(force_from:a [len: 8]) \
-	 map(force_from:b [len: 8])
-       {
-	 a[0] = 0;
-	 b[0] = 1;
-       }
-
-     Because both mappings have the force prefix, we know that they will be
-     allocated when calling the corresponding offloaded function, which means we
-     can mark the base pointers for a and b in the offloaded function as
-     restrict.  */
-
-  tree c;
-  for (c = clauses; c; c = OMP_CLAUSE_CHAIN (c))
-    {
-      if (OMP_CLAUSE_CODE (c) != OMP_CLAUSE_MAP)
-	return false;
-
-      switch (OMP_CLAUSE_MAP_KIND (c))
-	{
-	case GOMP_MAP_FORCE_ALLOC:
-	case GOMP_MAP_FORCE_TO:
-	case GOMP_MAP_FORCE_FROM:
-	case GOMP_MAP_FORCE_TOFROM:
-	  break;
-	default:
-	  return false;
-	}
-    }
-
-  return true;
-}
-
 /* Scan a GIMPLE_OMP_TARGET.  */
 
 static void
@@ -2349,20 +2283,13 @@ scan_omp_target (gomp_target *stmt, omp_context *outer_ctx)
   TYPE_NAME (ctx->record_type) = name;
   TYPE_ARTIFICIAL (ctx->record_type) = 1;
 
-  bool base_pointers_restrict = false;
   if (offloaded)
     {
       create_omp_child_function (ctx, false);
       gimple_omp_target_set_child_fn (stmt, ctx->cb.dst_fn);
-
-      base_pointers_restrict = omp_target_base_pointers_restrict_p (clauses);
-      if (base_pointers_restrict
-	  && dump_file && (dump_flags & TDF_DETAILS))
-	fprintf (dump_file,
-		 "Base pointers in offloaded function are restrict\n");
     }
 
-  scan_sharing_clauses (clauses, ctx, base_pointers_restrict);
+  scan_sharing_clauses (clauses, ctx);
   scan_omp (gimple_omp_body_ptr (stmt), ctx);
 
   if (TYPE_FIELDS (ctx->record_type) == NULL)
diff --git a/gcc/testsuite/ChangeLog b/gcc/testsuite/ChangeLog
index c30af52..1c06223 100644
--- a/gcc/testsuite/ChangeLog
+++ b/gcc/testsuite/ChangeLog
@@ -1,3 +1,36 @@
+2018-06-20  Chung-Lin Tang <cltang@codesourcery.com>
+	    Thomas Schwinge <thomas@codesourcery.com>
+	    Cesar Philippidis  <cesar@codesourcery.com>
+
+	* c-c++-common/goacc/declare-1.c: Update test case to utilize OpenACC
+	2.5 data clause semantics.
+	* c-c++-common/goacc/declare-2.c: Likewise.
+	* c-c++-common/goacc/default-4.c: Likewise.
+	* c-c++-common/goacc/finalize-1.c: New test.
+	* c-c++-common/goacc/kernels-alias-2.c: Update test case to utilize
+	OpenACC 2.5 data clause semantics.
+	* c-c++-common/goacc/kernels-alias.c: Likewise.
+	* c-c++-common/goacc/routine-5.c: Likewise.
+	* c-c++-common/goacc/update-if_present-1.c: New test.
+	* c-c++-common/goacc/update-if_present-2.c: New test.
+	* g++.dg/goacc/template.C: Update test case to utilize OpenACC
+	2.5 data clause semantics.
+	* gfortran.dg/goacc/combined-directives.f90: Likewise.
+	* gfortran.dg/goacc/data-tree.f95: Likewise.
+	* gfortran.dg/goacc/declare-2.f95: Likewise.
+	* gfortran.dg/goacc/default-4.f: Likewise.
+	* gfortran.dg/goacc/enter-exit-data.f95: Likewise.
+	* gfortran.dg/goacc/finalize-1.f: New test.
+	* gfortran.dg/goacc/kernels-alias-2.f95: Update test case to utilize
+	OpenACC 2.5 data clause semantics.
+	* gfortran.dg/goacc/kernels-alias.f95: Likewise.
+	* gfortran.dg/goacc/kernels-tree.f95: Likewise.
+	* gfortran.dg/goacc/nested-function-1.f90: Likewise.
+	* gfortran.dg/goacc/parallel-tree.f95: Likewise.
+	* gfortran.dg/goacc/reduction-promotions.f90: Likewise.
+	* gfortran.dg/goacc/update-if_present-1.f90: New test.
+	* gfortran.dg/goacc/update-if_present-2.f90: New test.
+
 2018-06-20  Jakub Jelinek  <jakub@redhat.com>
 
 	PR c++/86210
diff --git a/gcc/testsuite/c-c++-common/goacc/declare-1.c b/gcc/testsuite/c-c++-common/goacc/declare-1.c
index b036c63..35b1ccd 100644
--- a/gcc/testsuite/c-c++-common/goacc/declare-1.c
+++ b/gcc/testsuite/c-c++-common/goacc/declare-1.c
@@ -19,6 +19,12 @@ int v4;
 int v5, v6, v7, v8;
 #pragma acc declare create(v5, v6) copyin(v7, v8)
 
+int v9;
+#pragma acc declare present_or_copyin(v9)
+
+int v10;
+#pragma acc declare present_or_create(v10)
+
 void
 f (void)
 {
@@ -49,6 +55,12 @@ f (void)
   extern int ve4;
 #pragma acc declare link(ve4)
 
+  extern int ve5;
+#pragma acc declare present_or_copyin(ve5)
+ 
+  extern int ve6;
+#pragma acc declare present_or_create(ve6)
+
   int va5;
 #pragma acc declare copy(va5)
 
diff --git a/gcc/testsuite/c-c++-common/goacc/declare-2.c b/gcc/testsuite/c-c++-common/goacc/declare-2.c
index e41a0f5..33b8245 100644
--- a/gcc/testsuite/c-c++-common/goacc/declare-2.c
+++ b/gcc/testsuite/c-c++-common/goacc/declare-2.c
@@ -29,13 +29,7 @@ int v6;
 #pragma acc declare present_or_copy(v6) /* { dg-error "at file scope" } */
 
 int v7;
-#pragma acc declare present_or_copyin(v7) /* { dg-error "at file scope" } */
-
-int v8;
-#pragma acc declare present_or_copyout(v8) /* { dg-error "at file scope" } */
-
-int v9;
-#pragma acc declare present_or_create(v9) /* { dg-error "at file scope" } */
+#pragma acc declare present_or_copyout(v7) /* { dg-error "at file scope" } */
 
 int va10;
 #pragma acc declare create (va10)
@@ -67,13 +61,7 @@ f (void)
 #pragma acc declare present_or_copy(ve3) /* { dg-error "invalid use of" } */
 
   extern int ve4;
-#pragma acc declare present_or_copyin(ve4) /* { dg-error "invalid use of" } */
-
-  extern int ve5;
-#pragma acc declare present_or_copyout(ve5) /* { dg-error "invalid use of" } */
-
-  extern int ve6;
-#pragma acc declare present_or_create(ve6) /* { dg-error "invalid use of" } */
+#pragma acc declare present_or_copyout(ve4) /* { dg-error "invalid use of" } */
 
-#pragma acc declare present (v9) /* { dg-error "invalid use of" } */
+#pragma acc declare present (v2) /* { dg-error "invalid use of" } */
 }
diff --git a/gcc/testsuite/c-c++-common/goacc/default-4.c b/gcc/testsuite/c-c++-common/goacc/default-4.c
index dfa79bb..867175d 100644
--- a/gcc/testsuite/c-c++-common/goacc/default-4.c
+++ b/gcc/testsuite/c-c++-common/goacc/default-4.c
@@ -8,7 +8,7 @@ void f1 ()
   float f1_b[2];
 
 #pragma acc data copyin (f1_a) copyout (f1_b)
-  /* { dg-final { scan-tree-dump-times "omp target oacc_data map\\(force_from:f1_b \[^\\)\]+\\) map\\(force_to:f1_a" 1 "gimple" } } */
+  /* { dg-final { scan-tree-dump-times "omp target oacc_data map\\(from:f1_b \[^\\)\]+\\) map\\(to:f1_a" 1 "gimple" } } */
   {
 #pragma acc kernels
     /* { dg-final { scan-tree-dump-times "omp target oacc_kernels map\\(tofrom:f1_b \[^\\)\]+\\) map\\(tofrom:f1_a" 1 "gimple" } } */
@@ -29,7 +29,7 @@ void f2 ()
   float f2_b[2];
 
 #pragma acc data copyin (f2_a) copyout (f2_b)
-  /* { dg-final { scan-tree-dump-times "omp target oacc_data map\\(force_from:f2_b \[^\\)\]+\\) map\\(force_to:f2_a" 1 "gimple" } } */
+  /* { dg-final { scan-tree-dump-times "omp target oacc_data map\\(from:f2_b \[^\\)\]+\\) map\\(to:f2_a" 1 "gimple" } } */
   {
 #pragma acc kernels default (none)
     /* { dg-final { scan-tree-dump-times "omp target oacc_kernels default\\(none\\) map\\(tofrom:f2_b \[^\\)\]+\\) map\\(tofrom:f2_a" 1 "gimple" } } */
@@ -50,7 +50,7 @@ void f3 ()
   float f3_b[2];
 
 #pragma acc data copyin (f3_a) copyout (f3_b)
-  /* { dg-final { scan-tree-dump-times "omp target oacc_data map\\(force_from:f3_b \[^\\)\]+\\) map\\(force_to:f3_a" 1 "gimple" } } */
+  /* { dg-final { scan-tree-dump-times "omp target oacc_data map\\(from:f3_b \[^\\)\]+\\) map\\(to:f3_a" 1 "gimple" } } */
   {
 #pragma acc kernels default (present)
     /* { dg-final { scan-tree-dump-times "omp target oacc_kernels default\\(present\\) map\\(tofrom:f3_b \[^\\)\]+\\) map\\(tofrom:f3_a" 1 "gimple" } } */
diff --git a/gcc/testsuite/c-c++-common/goacc/finalize-1.c b/gcc/testsuite/c-c++-common/goacc/finalize-1.c
new file mode 100644
index 0000000..9482029
--- /dev/null
+++ b/gcc/testsuite/c-c++-common/goacc/finalize-1.c
@@ -0,0 +1,28 @@
+/* Test valid usage and processing of the finalize clause.  */
+
+/* { dg-additional-options "-fdump-tree-original -fdump-tree-gimple" } */
+
+extern int del_r;
+extern float del_f[3];
+extern double cpo_r[8];
+extern long cpo_f;
+
+void f ()
+{
+#pragma acc exit data delete (del_r)
+/* { dg-final { scan-tree-dump-times "(?n)#pragma acc exit data map\\(release:del_r\\);$" 1 "original" } }
+   { dg-final { scan-tree-dump-times "(?n)#pragma omp target oacc_enter_exit_data map\\(release:del_r \\\[len: \[0-9\]+\\\]\\)$" 1 "gimple" } } */
+
+#pragma acc exit data finalize delete (del_f)
+/* { dg-final { scan-tree-dump-times "(?n)#pragma acc exit data map\\(release:del_f\\) finalize;$" 1 "original" } }
+   { dg-final { scan-tree-dump-times "(?n)#pragma omp target oacc_enter_exit_data map\\(delete:del_f \\\[len: \[0-9\]+\\\]\\) finalize$" 1 "gimple" } } */
+
+#pragma acc exit data copyout (cpo_r)
+/* { dg-final { scan-tree-dump-times "(?n)#pragma acc exit data map\\(from:cpo_r\\);$" 1 "original" } }
+   { dg-final { scan-tree-dump-times "(?n)#pragma omp target oacc_enter_exit_data map\\(from:cpo_r \\\[len: \[0-9\]+\\\]\\)$" 1 "gimple" } } */
+
+#pragma acc exit data copyout (cpo_f) finalize
+/* { dg-final { scan-tree-dump-times "(?n)#pragma acc exit data finalize map\\(from:cpo_f\\);$" 1 "original" } }
+   { dg-final { scan-tree-dump-times "(?n)#pragma omp target oacc_enter_exit_data finalize map\\(force_from:cpo_f \\\[len: \[0-9\]+\\\]\\)$" 1 "gimple" } } */
+}
+
diff --git a/gcc/testsuite/c-c++-common/goacc/kernels-alias-2.c b/gcc/testsuite/c-c++-common/goacc/kernels-alias-2.c
index d437c47..7576a64 100644
--- a/gcc/testsuite/c-c++-common/goacc/kernels-alias-2.c
+++ b/gcc/testsuite/c-c++-common/goacc/kernels-alias-2.c
@@ -18,10 +18,12 @@ foo (void)
   }
 }
 
+/* The xfails occur due to the OpenACC 2.5 data semantics.  */
+
 /* { dg-final { scan-tree-dump-times "clique 1 base 1" 4 "ealias" } } */
-/* { dg-final { scan-tree-dump-times "clique 1 base 2" 1 "ealias" } } */
-/* { dg-final { scan-tree-dump-times "clique 1 base 3" 1 "ealias" } } */
-/* { dg-final { scan-tree-dump-times "clique 1 base 4" 1 "ealias" } } */
-/* { dg-final { scan-tree-dump-times "clique 1 base 5" 1 "ealias" } } */
+/* { dg-final { scan-tree-dump-times "clique 1 base 2" 1 "ealias" { xfail *-*-* } } } */
+/* { dg-final { scan-tree-dump-times "clique 1 base 3" 1 "ealias" { xfail *-*-* } } } */
+/* { dg-final { scan-tree-dump-times "clique 1 base 4" 1 "ealias" { xfail *-*-* } } } */
+/* { dg-final { scan-tree-dump-times "clique 1 base 5" 1 "ealias" { xfail *-*-* } } } */
 /* { dg-final { scan-tree-dump-times "(?n)clique .* base .*" 8 "ealias" } } */
 
diff --git a/gcc/testsuite/c-c++-common/goacc/kernels-alias.c b/gcc/testsuite/c-c++-common/goacc/kernels-alias.c
index 25821ab2..e8ff018 100644
--- a/gcc/testsuite/c-c++-common/goacc/kernels-alias.c
+++ b/gcc/testsuite/c-c++-common/goacc/kernels-alias.c
@@ -20,10 +20,12 @@ foo (void)
   }
 }
 
+/* The xfails occur due to the OpenACC 2.5 data semantics.  */
+
 /* { dg-final { scan-tree-dump-times "clique 1 base 1" 4 "ealias" } } */
-/* { dg-final { scan-tree-dump-times "clique 1 base 2" 1 "ealias" } } */
-/* { dg-final { scan-tree-dump-times "clique 1 base 3" 1 "ealias" } } */
-/* { dg-final { scan-tree-dump-times "clique 1 base 4" 1 "ealias" } } */
-/* { dg-final { scan-tree-dump-times "clique 1 base 5" 1 "ealias" } } */
+/* { dg-final { scan-tree-dump-times "clique 1 base 2" 1 "ealias" { xfail *-*-* } } } */
+/* { dg-final { scan-tree-dump-times "clique 1 base 3" 1 "ealias" { xfail *-*-* } } } */
+/* { dg-final { scan-tree-dump-times "clique 1 base 4" 1 "ealias" { xfail *-*-* } } } */
+/* { dg-final { scan-tree-dump-times "clique 1 base 5" 1 "ealias" { xfail *-*-* } } } */
 /* { dg-final { scan-tree-dump-times "(?n)clique .* base .*" 8 "ealias" } } */
 
diff --git a/gcc/testsuite/c-c++-common/goacc/routine-5.c b/gcc/testsuite/c-c++-common/goacc/routine-5.c
index b967a74..b759db3 100644
--- a/gcc/testsuite/c-c++-common/goacc/routine-5.c
+++ b/gcc/testsuite/c-c++-common/goacc/routine-5.c
@@ -4,11 +4,11 @@
 
 struct PC
 {
-#pragma acc routine /* { dg-error ".#pragma acc routine. must be at file scope" } */
+#pragma acc routine seq /* { dg-error ".#pragma acc routine. must be at file scope" } */
 };
 
 void PC1( /* { dg-bogus "variable or field .PC1. declared void" "TODO" { xfail c++ } } */
-#pragma acc routine
+#pragma acc routine seq
 	 /* { dg-error ".#pragma acc routine. must be at file scope" "" { target c } .-1 }
 	    { dg-error ".#pragma. is not allowed here" "" { target c++ } .-2 } */
 ) /* { dg-bogus "expected declaration specifiers or .\\.\\.\\.. before .\\). token" "TODO" { xfail c } } */
@@ -18,26 +18,26 @@ void PC1( /* { dg-bogus "variable or field .PC1. declared void" "TODO" { xfail c
 void PC2()
 {
   if (0)
-#pragma acc routine /* { dg-error ".#pragma acc routine. must be at file scope" } */
+#pragma acc routine seq /* { dg-error ".#pragma acc routine. must be at file scope" } */
     ;
 }
 
 void PC3()
 {
-#pragma acc routine /* { dg-error ".#pragma acc routine. must be at file scope" } */
+#pragma acc routine seq /* { dg-error ".#pragma acc routine. must be at file scope" } */
 }
 
 
 /* "( name )" syntax.  */
 
 #pragma acc routine ( /* { dg-error "expected (function name|unqualified-id) before end of line" } */
-#pragma acc routine () /* { dg-error "expected (function name|unqualified-id) before .\\). token" } */
-#pragma acc routine (+) /* { dg-error "expected (function name|unqualified-id) before .\\+. token" } */
-#pragma acc routine (?) /* { dg-error "expected (function name|unqualified-id) before .\\?. token" } */
-#pragma acc routine (:) /* { dg-error "expected (function name|unqualified-id) before .:. token" } */
-#pragma acc routine (4) /* { dg-error "expected (function name|unqualified-id) before numeric constant" } */
+#pragma acc routine () seq /* { dg-error "expected (function name|unqualified-id) before .\\). token" } */
+#pragma acc routine (+) seq /* { dg-error "expected (function name|unqualified-id) before .\\+. token" } */
+#pragma acc routine (?) seq /* { dg-error "expected (function name|unqualified-id) before .\\?. token" } */
+#pragma acc routine (:) seq /* { dg-error "expected (function name|unqualified-id) before .:. token" } */
+#pragma acc routine (4) seq /* { dg-error "expected (function name|unqualified-id) before numeric constant" } */
 #pragma acc routine ('4') /* { dg-error "expected (function name|unqualified-id) before .4." } */
-#pragma acc routine ("4") /* { dg-error "expected (function name|unqualified-id) before string constant" } */
+#pragma acc routine ("4") seq /* { dg-error "expected (function name|unqualified-id) before string constant" } */
 extern void R1(void);
 extern void R2(void);
 #pragma acc routine (R1, R2, R3) worker /* { dg-error "expected .\\). before .,. token" } */
@@ -49,84 +49,84 @@ extern void R2(void);
 /* "#pragma acc routine" not immediately followed by (a single) function
    declaration or definition.  */
 
-#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
+#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
 int a;
 
-#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by a single function declaration or definition" } */
+#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by a single function declaration or definition" } */
 void fn1 (void), fn1b (void);
 
-#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
+#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
 int b, fn2 (void);
 
-#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
+#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
 int b_, fn2_ (void), B_;
 
-#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by a single function declaration or definition" } */
+#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by a single function declaration or definition" } */
 int fn3 (void), b2;
 
-#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
+#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
 typedef struct c c;
 
-#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
+#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
 struct d {} d;
 
-#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
-#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by a single function declaration or definition" } */
+#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
+#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by a single function declaration or definition" } */
 void fn1_2 (void), fn1b_2 (void);
 
-#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
+#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
 #pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
 int b_2, fn2_2 (void);
 
-#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
-#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
+#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
+#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
 int b_2_, fn2_2_ (void), B_2_;
 
-#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
-#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by a single function declaration or definition" } */
+#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
+#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by a single function declaration or definition" } */
 int fn3_2 (void), b2_2;
 
-#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
-#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
+#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
+#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
 typedef struct c_2 c_2;
 
-#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
-#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
+#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
+#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
 struct d_2 {} d_2;
 
-#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
-#pragma acc routine
+#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
+#pragma acc routine seq
 int fn4 (void);
 
 int fn5a (void);
 int fn5b (void);
-#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
-#pragma acc routine (fn5a)
-#pragma acc routine (fn5b)
+#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
+#pragma acc routine (fn5a) seq
+#pragma acc routine (fn5b) seq
 int fn5 (void);
 
-#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
-#pragma acc routine (fn6a) /* { dg-error ".fn6a. has not been declared" } */
-#pragma acc routine (fn6b) /* { dg-error ".fn6b. has not been declared" } */
+#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
+#pragma acc routine (fn6a) seq /* { dg-error ".fn6a. has not been declared" } */
+#pragma acc routine (fn6b) seq /* { dg-error ".fn6b. has not been declared" } */
 int fn6 (void);
 
 #ifdef __cplusplus
 
-#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" "" { target c++ } } */
+#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" "" { target c++ } } */
 namespace f {}
 
 namespace g {}
 
-#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" "" { target c++ } } */
+#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" "" { target c++ } } */
 using namespace g;
 
-#pragma acc routine (g) /* { dg-error ".g. does not refer to a function" "" { target c++ } } */
+#pragma acc routine (g) seq /* { dg-error ".g. does not refer to a function" "" { target c++ } } */
 
 #endif /* __cplusplus */
 
-#pragma acc routine (a) /* { dg-error ".a. does not refer to a function" } */
+#pragma acc routine (a) seq /* { dg-error ".a. does not refer to a function" } */
   
-#pragma acc routine (c) /* { dg-error ".c. does not refer to a function" } */
+#pragma acc routine (c) seq /* { dg-error ".c. does not refer to a function" } */
 
 
 /* Static assert.  */
@@ -143,66 +143,24 @@ static_assert(0, ""); /* { dg-error "static assertion failed" "" { target c++11
 #endif
 void f_static_assert();
 /* Check that we already recognized "f_static_assert" as an OpenACC routine.  */
-#pragma acc routine (f_static_assert) /* { dg-error ".#pragma acc routine. already applied to .\[void \]*f_static_assert" "TODO" { xfail *-*-* } } */
+#pragma acc routine (f_static_assert) seq /* { dg-error ".#pragma acc routine. already applied to .\[void \]*f_static_assert" "TODO" { xfail *-*-* } } */
 
 
 /* __extension__ usage.  */
 
-#pragma acc routine
+#pragma acc routine seq
 __extension__ extern void ex1();
 #pragma acc routine (ex1) /* { dg-error ".#pragma acc routine. already applied to .\[void \]*ex1" } */
 
-#pragma acc routine
+#pragma acc routine seq
 __extension__ __extension__ __extension__ __extension__ __extension__ void ex2()
 {
 }
 #pragma acc routine (ex2) /* { dg-error ".#pragma acc routine. already applied to .\[void \]*ex2" } */
 
-#pragma acc routine /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
+#pragma acc routine seq /* { dg-error ".#pragma acc routine. not immediately followed by function declaration or definition" } */
 __extension__ int ex3;
-#pragma acc routine (ex3) /* { dg-error ".ex3. does not refer to a function" } */
-
-
-/* "#pragma acc routine" already applied.  */
-
-extern void fungsi_1();
-#pragma acc routine(fungsi_1) gang
-#pragma acc routine(fungsi_1) gang /* { dg-error ".#pragma acc routine. already applied to .\[void \]*fungsi_1" } */
-#pragma acc routine(fungsi_1) worker /* { dg-error ".#pragma acc routine. already applied to .\[void \]*fungsi_1" } */
-#pragma acc routine(fungsi_1) vector /* { dg-error ".#pragma acc routine. already applied to .\[void \]*fungsi_1" } */
-
-#pragma acc routine seq
-extern void fungsi_2();
-#pragma acc routine(fungsi_2) seq /* { dg-error ".#pragma acc routine. already applied to .\[void \]*fungsi_2." } */
-#pragma acc routine(fungsi_2) worker /* { dg-error ".#pragma acc routine. already applied to .\[void \]*fungsi_2." } */
-#pragma acc routine(fungsi_2) /* { dg-error ".#pragma acc routine. already applied to .\[void \]*fungsi_2." } */
-
-#pragma acc routine vector
-extern void fungsi_3();
-#pragma acc routine vector /* { dg-error ".#pragma acc routine. already applied to .\[void \]*fungsi_3." } */
-void fungsi_3()
-{
-}
-
-extern void fungsi_4();
-#pragma acc routine (fungsi_4) worker
-#pragma acc routine gang /* { dg-error ".#pragma acc routine. already applied to .\[void \]*fungsi_4." } */
-void fungsi_4()
-{
-}
-
-#pragma acc routine gang
-void fungsi_5()
-{
-}
-#pragma acc routine (fungsi_5) worker /* { dg-error ".#pragma acc routine. already applied to .\[void \]*fungsi_5." } */
-
-#pragma acc routine seq
-void fungsi_6()
-{
-}
-#pragma acc routine seq /* { dg-error ".#pragma acc routine. already applied to .\[void \]*fungsi_6." } */
-extern void fungsi_6();
+#pragma acc routine (ex3) seq /* { dg-error ".ex3. does not refer to a function" } */
 
 
 /* "#pragma acc routine" must be applied before.  */
@@ -214,11 +172,11 @@ void Foo ()
   Bar ();
 }
 
-#pragma acc routine (Bar) // { dg-error ".#pragma acc routine. must be applied before use" }
+#pragma acc routine (Bar) seq // { dg-error ".#pragma acc routine. must be applied before use" }
 
 #pragma acc routine (Foo) gang // { dg-error ".#pragma acc routine. must be applied before definition" }
 
-#pragma acc routine (Baz) // { dg-error "not been declared" }
+#pragma acc routine (Baz) seq // { dg-error "not been declared" }
 
 
 /* OpenACC declare.  */
@@ -227,7 +185,7 @@ int vb1;		/* { dg-error "directive for use" } */
 extern int vb2;		/* { dg-error "directive for use" } */
 static int vb3;		/* { dg-error "directive for use" } */
 
-#pragma acc routine
+#pragma acc routine seq
 int
 func1 (int a)
 {
@@ -238,7 +196,7 @@ func1 (int a)
   return vb3;
 }
 
-#pragma acc routine
+#pragma acc routine seq
 int
 func2 (int a)
 {
@@ -256,7 +214,7 @@ extern int vb6;			/* { dg-error "clause used in" } */
 static int vb7;			/* { dg-error "clause used in" } */
 #pragma acc declare link (vb7)
 
-#pragma acc routine
+#pragma acc routine seq
 int
 func3 (int a)
 {
@@ -273,7 +231,7 @@ extern int vb9;
 static int vb10;
 #pragma acc declare create (vb10)
 
-#pragma acc routine
+#pragma acc routine seq
 int
 func4 (int a)
 {
@@ -291,7 +249,7 @@ extern int vb12;
 extern int vb13;
 #pragma acc declare device_resident (vb13)
 
-#pragma acc routine
+#pragma acc routine seq
 int
 func5 (int a)
 {
@@ -302,7 +260,7 @@ func5 (int a)
   return vb13;
 }
 
-#pragma acc routine
+#pragma acc routine seq
 int
 func6 (int a)
 {
diff --git a/gcc/testsuite/c-c++-common/goacc/update-if_present-1.c b/gcc/testsuite/c-c++-common/goacc/update-if_present-1.c
new file mode 100644
index 0000000..c34a0e4
--- /dev/null
+++ b/gcc/testsuite/c-c++-common/goacc/update-if_present-1.c
@@ -0,0 +1,28 @@
+/* Test valid usages of the if_present clause.  */
+
+/* { dg-additional-options "-fdump-tree-omplower" } */
+
+void
+t ()
+{
+  int a, b, c[10];
+
+#pragma acc update self(a) if_present
+#pragma acc update device(b) async if_present
+#pragma acc update host(c[1:3]) wait(4) if_present
+#pragma acc update self(c) device(b) host (a) async(10) if (a == 5) if_present
+
+#pragma acc update self(a)
+#pragma acc update device(b) async
+#pragma acc update host(c[1:3]) wait(4)
+#pragma acc update self(c) device(b) host (a) async(10) if (a == 5)
+}
+
+/* { dg-final { scan-tree-dump-times "omp target oacc_update if_present map.from:a .len: 4.." 1 "omplower" } } */
+/* { dg-final { scan-tree-dump-times "omp target oacc_update if_present async.-1. map.to:b .len: 4.." 1 "omplower" } } */
+/* { dg-final { scan-tree-dump-times "omp target oacc_update if_present wait.4. map.from:c.1. .len: 12.." 1 "omplower" } } */
+/* { dg-final { scan-tree-dump-times "omp target oacc_update if_present if.... async.10. map.from:a .len: 4.. map.to:b .len: 4.. map.from:c .len: 40.." 1 "omplower" } } */
+/* { dg-final { scan-tree-dump-times "omp target oacc_update map.force_from:a .len: 4.." 1 "omplower" } } */
+/* { dg-final { scan-tree-dump-times "omp target oacc_update async.-1. map.force_to:b .len: 4.." 1 "omplower" } } */
+/* { dg-final { scan-tree-dump-times "omp target oacc_update wait.4. map.force_from:c.1. .len: 12.." 1 "omplower" } } */
+/* { dg-final { scan-tree-dump-times "omp target oacc_update if.... async.10. map.force_from:a .len: 4.. map.force_to:b .len: 4.. map.force_from:c .len: 40.." 1 "omplower" } } */
diff --git a/gcc/testsuite/c-c++-common/goacc/update-if_present-2.c b/gcc/testsuite/c-c++-common/goacc/update-if_present-2.c
new file mode 100644
index 0000000..974f1b8
--- /dev/null
+++ b/gcc/testsuite/c-c++-common/goacc/update-if_present-2.c
@@ -0,0 +1,42 @@
+/* Test invalid usages of the if_present clause.  */
+
+#pragma acc routine gang if_present /* { dg-error "'if_present' is not valid" } */
+void
+t1 ()
+{
+  int a, b, c[10];
+
+#pragma acc enter data copyin(a) if_present /* { dg-error "'if_present' is not valid" } */
+#pragma acc exit data copyout(a) if_present /* { dg-error "'if_present' is not valid" } */
+
+#pragma acc data copy(a) if_present /* { dg-error "'if_present' is not valid" } */
+  {
+  }
+
+#pragma acc declare create(c) if_present /* { dg-error "'if_present' is not valid" } */
+
+#pragma acc init if_present
+#pragma acc shutdown if_present
+}
+
+void
+t2 ()
+{
+  int a, b, c[10];
+
+#pragma acc update self(a)
+#pragma acc parallel
+#pragma acc loop if_present /* { dg-error "'if_present' is not valid" } */
+  for (b = 1; b < 10; b++)
+    ;
+#pragma acc end parallel
+
+#pragma acc kernels loop if_present /* { dg-error "'if_present' is not valid" } */
+  for (b = 1; b < 10; b++)
+    ;
+
+#pragma acc parallel loop if_present /* { dg-error "'if_present' is not valid" } */
+    for (b = 1; b < 10; b++)
+    ;
+}
+
diff --git a/gcc/testsuite/g++.dg/goacc/template.C b/gcc/testsuite/g++.dg/goacc/template.C
index 852f42f..dae92b0 100644
--- a/gcc/testsuite/g++.dg/goacc/template.C
+++ b/gcc/testsuite/g++.dg/goacc/template.C
@@ -1,4 +1,4 @@
-#pragma acc routine
+#pragma acc routine seq
 template <typename T> T
 accDouble(int val)
 {
@@ -31,7 +31,7 @@ oacc_parallel_copy (T a)
 
 #pragma acc parallel num_gangs (a) if (1)
   {
-#pragma acc loop independent collapse (2) gang
+#pragma acc loop independent collapse (2)
     for (int i = 0; i < a; i++)
       for (int j = 0; j < 5; j++)
 	b = a;
@@ -86,6 +86,8 @@ oacc_parallel_copy (T a)
 #pragma acc update self (b)
 #pragma acc update device (b)
 #pragma acc exit data delete (b)
+#pragma acc exit data finalize copyout (b)
+#pragma acc exit data delete (b) finalize
 
   return b;
 }
@@ -133,6 +135,13 @@ oacc_kernels_copy (T a)
     b = a;
   }
 
+#pragma acc update host (b)
+#pragma acc update self (b)
+#pragma acc update device (b)
+#pragma acc exit data delete (b)
+#pragma acc exit data finalize copyout (b)
+#pragma acc exit data delete (b) finalize
+
   return b;
 }
 
diff --git a/gcc/testsuite/gfortran.dg/goacc/combined-directives.f90 b/gcc/testsuite/gfortran.dg/goacc/combined-directives.f90
index 42a447a..9563492 100644
--- a/gcc/testsuite/gfortran.dg/goacc/combined-directives.f90
+++ b/gcc/testsuite/gfortran.dg/goacc/combined-directives.f90
@@ -146,5 +146,5 @@ end subroutine test
 ! { dg-final { scan-tree-dump-times "acc loop private.i. private.j. tile.2, 3" 2 "gimple" } }
 ! { dg-final { scan-tree-dump-times "acc loop private.i. independent" 2 "gimple" } }
 ! { dg-final { scan-tree-dump-times "private.z" 2 "gimple" } }
-! { dg-final { scan-tree-dump-times "omp target oacc_\[^ \]+ map.force_tofrom:y" 2 "gimple" } }
+! { dg-final { scan-tree-dump-times "omp target oacc_\[^ \]+ map.tofrom:y" 2 "gimple" } }
 ! { dg-final { scan-tree-dump-times "acc loop private.i. reduction..:y." 2 "gimple" } }
diff --git a/gcc/testsuite/gfortran.dg/goacc/data-tree.f95 b/gcc/testsuite/gfortran.dg/goacc/data-tree.f95
index 44efc8a..f16d62c 100644
--- a/gcc/testsuite/gfortran.dg/goacc/data-tree.f95
+++ b/gcc/testsuite/gfortran.dg/goacc/data-tree.f95
@@ -15,10 +15,10 @@ end program test
 ! { dg-final { scan-tree-dump-times "pragma acc data" 1 "original" } } 
 
 ! { dg-final { scan-tree-dump-times "if" 1 "original" } }
-! { dg-final { scan-tree-dump-times "map\\(force_tofrom:i\\)" 1 "original" } } 
-! { dg-final { scan-tree-dump-times "map\\(force_to:j\\)" 1 "original" } } 
-! { dg-final { scan-tree-dump-times "map\\(force_from:k\\)" 1 "original" } } 
-! { dg-final { scan-tree-dump-times "map\\(force_alloc:m\\)" 1 "original" } } 
+! { dg-final { scan-tree-dump-times "map\\(tofrom:i\\)" 1 "original" } } 
+! { dg-final { scan-tree-dump-times "map\\(to:j\\)" 1 "original" } } 
+! { dg-final { scan-tree-dump-times "map\\(from:k\\)" 1 "original" } } 
+! { dg-final { scan-tree-dump-times "map\\(alloc:m\\)" 1 "original" } } 
 
 ! { dg-final { scan-tree-dump-times "map\\(force_present:o\\)" 1 "original" } } 
 ! { dg-final { scan-tree-dump-times "map\\(tofrom:p\\)" 1 "original" } } 
diff --git a/gcc/testsuite/gfortran.dg/goacc/declare-2.f95 b/gcc/testsuite/gfortran.dg/goacc/declare-2.f95
index aa1704f..7aa3dab 100644
--- a/gcc/testsuite/gfortran.dg/goacc/declare-2.f95
+++ b/gcc/testsuite/gfortran.dg/goacc/declare-2.f95
@@ -11,11 +11,11 @@ subroutine asubr (b)
   !$acc declare copyout (b) ! { dg-error "Invalid clause in module" }
   !$acc declare present (b) ! { dg-error "Invalid clause in module" }
   !$acc declare present_or_copy (b) ! { dg-error "Invalid clause in module" }
-  !$acc declare present_or_copyin (b) ! { dg-error "Invalid clause in module" }
+  !$acc declare present_or_copyin (b) ! { dg-error "present on multiple" }
   !$acc declare present_or_copyout (b) ! { dg-error "Invalid clause in module" }
-  !$acc declare present_or_create (b) ! { dg-error "Invalid clause in module" }
+  !$acc declare present_or_create (b) ! { dg-error "present on multiple" }
   !$acc declare deviceptr (b) ! { dg-error "Invalid clause in module" }
-  !$acc declare create (b) copyin (b) ! { dg-error "present on multiple clauses" }
+  !$acc declare create (b) copyin (b) ! { dg-error "present on multiple" }
 
 end subroutine
 
diff --git a/gcc/testsuite/gfortran.dg/goacc/default-4.f b/gcc/testsuite/gfortran.dg/goacc/default-4.f
index 77291f4..30f411f 100644
--- a/gcc/testsuite/gfortran.dg/goacc/default-4.f
+++ b/gcc/testsuite/gfortran.dg/goacc/default-4.f
@@ -8,7 +8,7 @@
       REAL, DIMENSION (2) :: F1_B
 
 !$ACC DATA COPYIN (F1_A) COPYOUT (F1_B)
-! { dg-final { scan-tree-dump-times "omp target oacc_data map\\(force_to:f1_a \[^\\)\]+\\) map\\(force_from:f1_b" 1 "gimple" } }
+! { dg-final { scan-tree-dump-times "omp target oacc_data map\\(to:f1_a \[^\\)\]+\\) map\\(from:f1_b" 1 "gimple" } }
 !$ACC KERNELS
 ! { dg-final { scan-tree-dump-times "omp target oacc_kernels map\\(tofrom:f1_b \[^\\)\]+\\) map\\(tofrom:f1_a" 1 "gimple" } }
       F1_B(1) = F1_A;
@@ -26,7 +26,7 @@
       REAL, DIMENSION (2) :: F2_B
 
 !$ACC DATA COPYIN (F2_A) COPYOUT (F2_B)
-! { dg-final { scan-tree-dump-times "omp target oacc_data map\\(force_to:f2_a \[^\\)\]+\\) map\\(force_from:f2_b" 1 "gimple" } }
+! { dg-final { scan-tree-dump-times "omp target oacc_data map\\(to:f2_a \[^\\)\]+\\) map\\(from:f2_b" 1 "gimple" } }
 !$ACC KERNELS DEFAULT (NONE)
 ! { dg-final { scan-tree-dump-times "omp target oacc_kernels default\\(none\\) map\\(tofrom:f2_b \[^\\)\]+\\) map\\(tofrom:f2_a" 1 "gimple" } }
       F2_B(1) = F2_A;
@@ -44,7 +44,7 @@
       REAL, DIMENSION (2) :: F3_B
 
 !$ACC DATA COPYIN (F3_A) COPYOUT (F3_B)
-! { dg-final { scan-tree-dump-times "omp target oacc_data map\\(force_to:f3_a \[^\\)\]+\\) map\\(force_from:f3_b" 1 "gimple" } }
+! { dg-final { scan-tree-dump-times "omp target oacc_data map\\(to:f3_a \[^\\)\]+\\) map\\(from:f3_b" 1 "gimple" } }
 !$ACC KERNELS DEFAULT (PRESENT)
 ! { dg-final { scan-tree-dump-times "omp target oacc_kernels default\\(present\\) map\\(tofrom:f3_b \[^\\)\]+\\) map\\(tofrom:f3_a" 1 "gimple" } }
       F3_B(1) = F3_A;
diff --git a/gcc/testsuite/gfortran.dg/goacc/enter-exit-data.f95 b/gcc/testsuite/gfortran.dg/goacc/enter-exit-data.f95
index 8f1715e..805459c 100644
--- a/gcc/testsuite/gfortran.dg/goacc/enter-exit-data.f95
+++ b/gcc/testsuite/gfortran.dg/goacc/enter-exit-data.f95
@@ -84,5 +84,8 @@ contains
   !$acc exit data delete (tip) ! { dg-error "POINTER" }
   !$acc exit data delete (tia) ! { dg-error "ALLOCATABLE" }
   !$acc exit data copyout (i) delete (i) ! { dg-error "multiple clauses" }
+  !$acc exit data finalize
+  !$acc exit data finalize copyout (i)
+  !$acc exit data finalize delete (i)
   end subroutine foo
 end module test
diff --git a/gcc/testsuite/gfortran.dg/goacc/finalize-1.f b/gcc/testsuite/gfortran.dg/goacc/finalize-1.f
new file mode 100644
index 0000000..5c7a921
--- /dev/null
+++ b/gcc/testsuite/gfortran.dg/goacc/finalize-1.f
@@ -0,0 +1,27 @@
+! Test valid usage and processing of the finalize clause.
+
+! { dg-additional-options "-fdump-tree-original -fdump-tree-gimple" }
+
+      SUBROUTINE f
+      IMPLICIT NONE
+      INTEGER :: del_r
+      REAL, DIMENSION (3) :: del_f
+      DOUBLE PRECISION, DIMENSION (8) :: cpo_r
+      LOGICAL :: cpo_f
+
+!$ACC EXIT DATA DELETE (del_r)
+! { dg-final { scan-tree-dump-times "(?n)#pragma acc exit data map\\(release:del_r\\);$" 1 "original" } }
+! { dg-final { scan-tree-dump-times "(?n)#pragma omp target oacc_enter_exit_data map\\(release:del_r \\\[len: \[0-9\]+\\\]\\)$" 1 "gimple" } }
+
+!$ACC EXIT DATA FINALIZE DELETE (del_f)
+! { dg-final { scan-tree-dump-times "(?n)#pragma acc exit data map\\(release:del_f\\) finalize;$" 1 "original" } }
+! { dg-final { scan-tree-dump-times "(?n)#pragma omp target oacc_enter_exit_data map\\(delete:del_f \\\[len: \[0-9\]+\\\]\\) finalize$" 1 "gimple" } }
+
+!$ACC EXIT DATA COPYOUT (cpo_r)
+! { dg-final { scan-tree-dump-times "(?n)#pragma acc exit data map\\(from:cpo_r\\);$" 1 "original" } }
+! { dg-final { scan-tree-dump-times "(?n)#pragma omp target oacc_enter_exit_data map\\(from:cpo_r \\\[len: \[0-9\]+\\\]\\)$" 1 "gimple" } }
+
+!$ACC EXIT DATA COPYOUT (cpo_f) FINALIZE
+! { dg-final { scan-tree-dump-times "(?n)#pragma acc exit data map\\(from:cpo_f\\) finalize;$" 1 "original" } }
+! { dg-final { scan-tree-dump-times "(?n)#pragma omp target oacc_enter_exit_data map\\(force_from:cpo_f \\\[len: \[0-9\]+\\\]\\) finalize$" 1 "gimple" } }
+      END SUBROUTINE f
diff --git a/gcc/testsuite/gfortran.dg/goacc/kernels-alias-2.f95 b/gcc/testsuite/gfortran.dg/goacc/kernels-alias-2.f95
index 7e348dd..6a9f241 100644
--- a/gcc/testsuite/gfortran.dg/goacc/kernels-alias-2.f95
+++ b/gcc/testsuite/gfortran.dg/goacc/kernels-alias-2.f95
@@ -15,9 +15,11 @@ program main
 
 end program main
 
+! The xfails occur in light of the new OpenACC data semantics.
+
 ! { dg-final { scan-tree-dump-times "clique 1 base 1" 4 "ealias" } }
-! { dg-final { scan-tree-dump-times "clique 1 base 2" 1 "ealias" } }
-! { dg-final { scan-tree-dump-times "clique 1 base 3" 1 "ealias" } }
-! { dg-final { scan-tree-dump-times "clique 1 base 4" 1 "ealias" } }
-! { dg-final { scan-tree-dump-times "clique 1 base 5" 1 "ealias" } }
+! { dg-final { scan-tree-dump-times "clique 1 base 2" 1 "ealias" { xfail *-*-* } } }
+! { dg-final { scan-tree-dump-times "clique 1 base 3" 1 "ealias" { xfail *-*-* } } }
+! { dg-final { scan-tree-dump-times "clique 1 base 4" 1 "ealias" { xfail *-*-* } } }
+! { dg-final { scan-tree-dump-times "clique 1 base 5" 1 "ealias" { xfail *-*-* } } }
 ! { dg-final { scan-tree-dump-times "(?n)clique .* base .*" 8 "ealias" } }
diff --git a/gcc/testsuite/gfortran.dg/goacc/kernels-alias.f95 b/gcc/testsuite/gfortran.dg/goacc/kernels-alias.f95
index 8d6ccb3..62f9a71 100644
--- a/gcc/testsuite/gfortran.dg/goacc/kernels-alias.f95
+++ b/gcc/testsuite/gfortran.dg/goacc/kernels-alias.f95
@@ -15,9 +15,11 @@ program main
 
 end program main
 
+! The xfails occur in light of the new OpenACC data semantics.
+
 ! { dg-final { scan-tree-dump-times "clique 1 base 1" 4 "ealias" } }
-! { dg-final { scan-tree-dump-times "clique 1 base 2" 1 "ealias" } }
-! { dg-final { scan-tree-dump-times "clique 1 base 3" 1 "ealias" } }
-! { dg-final { scan-tree-dump-times "clique 1 base 4" 1 "ealias" } }
-! { dg-final { scan-tree-dump-times "clique 1 base 5" 1 "ealias" } }
+! { dg-final { scan-tree-dump-times "clique 1 base 2" 1 "ealias" { xfail *-*-* } } }
+! { dg-final { scan-tree-dump-times "clique 1 base 3" 1 "ealias" { xfail *-*-* } } }
+! { dg-final { scan-tree-dump-times "clique 1 base 4" 1 "ealias" { xfail *-*-* } } }
+! { dg-final { scan-tree-dump-times "clique 1 base 5" 1 "ealias" { xfail *-*-* } } }
 ! { dg-final { scan-tree-dump-times "(?n)clique .* base .*" 8 "ealias" } }
diff --git a/gcc/testsuite/gfortran.dg/goacc/kernels-tree.f95 b/gcc/testsuite/gfortran.dg/goacc/kernels-tree.f95
index 7daca59..a70f1e7 100644
--- a/gcc/testsuite/gfortran.dg/goacc/kernels-tree.f95
+++ b/gcc/testsuite/gfortran.dg/goacc/kernels-tree.f95
@@ -21,10 +21,10 @@ end program test
 ! { dg-final { scan-tree-dump-times "num_workers" 1 "original" } } 
 ! { dg-final { scan-tree-dump-times "vector_length" 1 "original" } } 
 
-! { dg-final { scan-tree-dump-times "map\\(force_tofrom:i\\)" 1 "original" } } 
-! { dg-final { scan-tree-dump-times "map\\(force_to:j\\)" 1 "original" } } 
-! { dg-final { scan-tree-dump-times "map\\(force_from:k\\)" 1 "original" } } 
-! { dg-final { scan-tree-dump-times "map\\(force_alloc:m\\)" 1 "original" } } 
+! { dg-final { scan-tree-dump-times "map\\(tofrom:i\\)" 1 "original" } } 
+! { dg-final { scan-tree-dump-times "map\\(to:j\\)" 1 "original" } } 
+! { dg-final { scan-tree-dump-times "map\\(from:k\\)" 1 "original" } } 
+! { dg-final { scan-tree-dump-times "map\\(alloc:m\\)" 1 "original" } } 
 
 ! { dg-final { scan-tree-dump-times "map\\(force_present:o\\)" 1 "original" } } 
 ! { dg-final { scan-tree-dump-times "map\\(tofrom:p\\)" 1 "original" } } 
diff --git a/gcc/testsuite/gfortran.dg/goacc/nested-function-1.f90 b/gcc/testsuite/gfortran.dg/goacc/nested-function-1.f90
index 2fcaa40..005193f 100644
--- a/gcc/testsuite/gfortran.dg/goacc/nested-function-1.f90
+++ b/gcc/testsuite/gfortran.dg/goacc/nested-function-1.f90
@@ -25,6 +25,8 @@ contains
     local_a (:) = 5
     local_arg = 5
 
+    !$acc update device(local_a) if_present
+
     !$acc kernels loop &
     !$acc gang(num:local_arg) worker(local_arg) vector(local_arg) &
     !$acc wait async(local_arg)
@@ -54,12 +56,16 @@ contains
        enddo
     enddo
     !$acc end kernels loop
+
+    !$acc exit data copyout(local_a) delete(local_i) finalize
   end subroutine local
 
   subroutine nonlocal ()
     nonlocal_a (:) = 5
     nonlocal_arg = 5
 
+    !$acc update device(nonlocal_a) if_present
+
     !$acc kernels loop &
     !$acc gang(num:nonlocal_arg) worker(nonlocal_arg) vector(nonlocal_arg) &
     !$acc wait async(nonlocal_arg)
@@ -89,5 +95,7 @@ contains
        enddo
     enddo
     !$acc end kernels loop
+
+    !$acc exit data copyout(nonlocal_a) delete(nonlocal_i) finalize
   end subroutine nonlocal
 end program main
diff --git a/gcc/testsuite/gfortran.dg/goacc/parallel-tree.f95 b/gcc/testsuite/gfortran.dg/goacc/parallel-tree.f95
index 5b2e01d..2697bb7 100644
--- a/gcc/testsuite/gfortran.dg/goacc/parallel-tree.f95
+++ b/gcc/testsuite/gfortran.dg/goacc/parallel-tree.f95
@@ -1,5 +1,4 @@
-! { dg-do compile } 
-! { dg-additional-options "-fdump-tree-original" } 
+! { dg-additional-options "-fdump-tree-original" }
 
 ! test for tree-dump-original and spaces-commas
 
@@ -15,6 +14,7 @@ program test
   !$acc end parallel
 
 end program test
+
 ! { dg-final { scan-tree-dump-times "pragma acc parallel" 1 "original" } } 
 
 ! { dg-final { scan-tree-dump-times "if" 1 "original" } }
@@ -24,10 +24,10 @@ end program test
 ! { dg-final { scan-tree-dump-times "vector_length" 1 "original" } } 
 
 ! { dg-final { scan-tree-dump-times "reduction\\(max:q\\)" 1 "original" } } 
-! { dg-final { scan-tree-dump-times "map\\(force_tofrom:i\\)" 1 "original" } } 
-! { dg-final { scan-tree-dump-times "map\\(force_to:j\\)" 1 "original" } } 
-! { dg-final { scan-tree-dump-times "map\\(force_from:k\\)" 1 "original" } } 
-! { dg-final { scan-tree-dump-times "map\\(force_alloc:m\\)" 1 "original" } } 
+! { dg-final { scan-tree-dump-times "map\\(tofrom:i\\)" 1 "original" } } 
+! { dg-final { scan-tree-dump-times "map\\(to:j\\)" 1 "original" } } 
+! { dg-final { scan-tree-dump-times "map\\(from:k\\)" 1 "original" } } 
+! { dg-final { scan-tree-dump-times "map\\(alloc:m\\)" 1 "original" } } 
 
 ! { dg-final { scan-tree-dump-times "map\\(force_present:o\\)" 1 "original" } } 
 ! { dg-final { scan-tree-dump-times "map\\(tofrom:p\\)" 1 "original" } } 
diff --git a/gcc/testsuite/gfortran.dg/goacc/reduction-promotions.f90 b/gcc/testsuite/gfortran.dg/goacc/reduction-promotions.f90
index 6ff913a..1d247ca 100644
--- a/gcc/testsuite/gfortran.dg/goacc/reduction-promotions.f90
+++ b/gcc/testsuite/gfortran.dg/goacc/reduction-promotions.f90
@@ -38,9 +38,7 @@ program test
   !$acc end parallel
 end program test
 
-! { dg-final { scan-tree-dump-times "map.tofrom:v1" 8 "gimple" } }
-! { dg-final { scan-tree-dump-times "map.tofrom:v2" 8 "gimple" } }
-! { dg-final { scan-tree-dump-times "map.force_tofrom:v1" 1 "gimple" } }
-! { dg-final { scan-tree-dump-times "map.force_tofrom:v2" 1 "gimple" } }
+! { dg-final { scan-tree-dump-times "map.tofrom:v1" 9 "gimple" } }
+! { dg-final { scan-tree-dump-times "map.tofrom:v2" 9 "gimple" } }
 ! { dg-final { scan-tree-dump-times "map.force_present:v1" 1 "gimple" } }
 ! { dg-final { scan-tree-dump-times "map.force_present:v2" 1 "gimple" } }
diff --git a/gcc/testsuite/gfortran.dg/goacc/update-if_present-1.f90 b/gcc/testsuite/gfortran.dg/goacc/update-if_present-1.f90
new file mode 100644
index 0000000..a183aae
--- /dev/null
+++ b/gcc/testsuite/gfortran.dg/goacc/update-if_present-1.f90
@@ -0,0 +1,27 @@
+! Test valid usages of the if_present clause.
+
+! { dg-additional-options "-fdump-tree-omplower" }
+
+subroutine t
+  implicit none
+  integer a, b, c(10)
+  real, allocatable :: x, y, z(:)
+  
+  a = 5
+  b = 10
+  c(:) = -1
+
+  allocate (x, y, z(100))
+  
+  !$acc update self(a) if_present
+  !$acc update device(b) if_present async
+  !$acc update host(c(1:3)) wait(4) if_present
+  !$acc update self(c) device(a) host(b) if_present async(10) if(a == 10) 
+
+  !$acc update self(x) if_present
+  !$acc update device(y) if_present async
+  !$acc update host(z(1:3)) wait(3) if_present
+  !$acc update self(z) device(y) host(x) if_present async(4) if(a == 1) 
+end subroutine t
+
+! { dg-final { scan-tree-dump-times " if_present" 8 "omplower" } }
diff --git a/gcc/testsuite/gfortran.dg/goacc/update-if_present-2.f90 b/gcc/testsuite/gfortran.dg/goacc/update-if_present-2.f90
new file mode 100644
index 0000000..e73c2dc
--- /dev/null
+++ b/gcc/testsuite/gfortran.dg/goacc/update-if_present-2.f90
@@ -0,0 +1,52 @@
+! Test invalid usages of the if_present clause.
+
+subroutine t1
+  implicit none
+  !$acc routine gang if_present ! { dg-error "Unclassifiable OpenACC directive" }
+  integer a, b, c(10)
+  real, allocatable :: x, y, z(:)
+
+  a = 5
+  b = 10
+  c(:) = -1
+
+  allocate (x, y, z(100))
+
+  !$acc enter data copyin(a) if_present ! { dg-error "Unclassifiable OpenACC directive" }
+  !$acc exit data copyout(a) if_present ! { dg-error "Unclassifiable OpenACC directive" }
+
+  !$acc data copy(a) if_present ! { dg-error "Unclassifiable OpenACC directive" }
+  !$acc end data ! { dg-error "Unexpected ..ACC END DATA statement" }
+
+  !$acc declare link(a) if_present ! { dg-error "Unexpected junk after" }
+
+  !$acc init if_present ! { dg-error "Unclassifiable OpenACC directive" }
+  !$acc shutdown if_present ! { dg-error "Unclassifiable OpenACC directive" }
+  
+  !$acc update self(a) device_type(nvidia) device(b) if_present ! { dg-error "Unclassifiable OpenACC directive" }
+end subroutine t1
+
+subroutine t2
+  implicit none
+  integer a, b, c(10)
+
+  a = 5
+  b = 10
+  c(:) = -1
+
+  !$acc parallel
+  !$acc loop if_present ! { dg-error "Unclassifiable OpenACC directive" }
+  do b = 1, 10
+  end do
+  !$acc end parallel
+
+  !$acc kernels loop if_present ! { dg-error "Unclassifiable OpenACC directive" }
+  do b = 1, 10
+  end do
+  !$acc end kernels loop ! { dg-error "Unexpected ..ACC END KERNELS LOOP statement" }
+
+  !$acc parallel loop if_present ! { dg-error "Unclassifiable OpenACC directive" }
+  do b = 1, 10
+  end do
+  !$acc end parallel loop   ! { dg-error "Unexpected ..ACC END PARALLEL LOOP statement" }
+end subroutine t2
diff --git a/gcc/tree-core.h b/gcc/tree-core.h
index 2bebb22..4a04e9e 100644
--- a/gcc/tree-core.h
+++ b/gcc/tree-core.h
@@ -454,7 +454,13 @@ enum omp_clause_code {
 
   /* OpenMP internal-only clause to specify grid dimensions of a gridified
      kernel.  */
-  OMP_CLAUSE__GRIDDIM_
+  OMP_CLAUSE__GRIDDIM_,
+
+  /* OpenACC clause: if_present.  */
+  OMP_CLAUSE_IF_PRESENT,
+
+  /* OpenACC clause: finalize.  */
+  OMP_CLAUSE_FINALIZE
 };
 
 #undef DEFTREESTRUCT
diff --git a/gcc/tree-nested.c b/gcc/tree-nested.c
index b335d6b..257ceae 100644
--- a/gcc/tree-nested.c
+++ b/gcc/tree-nested.c
@@ -1333,6 +1333,8 @@ convert_nonlocal_omp_clauses (tree *pclauses, struct walk_stmt_info *wi)
 	case OMP_CLAUSE_SEQ:
 	case OMP_CLAUSE_INDEPENDENT:
 	case OMP_CLAUSE_AUTO:
+	case OMP_CLAUSE_IF_PRESENT:
+	case OMP_CLAUSE_FINALIZE:
 	  break;
 
 	  /* The following clause belongs to the OpenACC cache directive, which
@@ -2022,6 +2024,8 @@ convert_local_omp_clauses (tree *pclauses, struct walk_stmt_info *wi)
 	case OMP_CLAUSE_SEQ:
 	case OMP_CLAUSE_INDEPENDENT:
 	case OMP_CLAUSE_AUTO:
+	case OMP_CLAUSE_IF_PRESENT:
+	case OMP_CLAUSE_FINALIZE:
 	  break;
 
 	  /* The following clause belongs to the OpenACC cache directive, which
diff --git a/gcc/tree-pretty-print.c b/gcc/tree-pretty-print.c
index 63ec823..e65c40a 100644
--- a/gcc/tree-pretty-print.c
+++ b/gcc/tree-pretty-print.c
@@ -1045,6 +1045,12 @@ dump_omp_clause (pretty_printer *pp, tree clause, int spc, dump_flags_t flags)
 			 false);
       pp_right_paren (pp);
       break;
+    case OMP_CLAUSE_IF_PRESENT:
+      pp_string (pp, "if_present");
+      break;
+    case OMP_CLAUSE_FINALIZE:
+      pp_string (pp, "finalize");
+      break;
 
     default:
       /* Should never happen.  */
diff --git a/gcc/tree.c b/gcc/tree.c
index 3f75f7f..608ca7e 100644
--- a/gcc/tree.c
+++ b/gcc/tree.c
@@ -343,6 +343,8 @@ unsigned const char omp_clause_num_ops[] =
   1, /* OMP_CLAUSE_VECTOR_LENGTH  */
   3, /* OMP_CLAUSE_TILE  */
   2, /* OMP_CLAUSE__GRIDDIM_  */
+  0, /* OMP_CLAUSE_IF_PRESENT */
+  0, /* OMP_CLAUSE_FINALIZE */
 };
 
 const char * const omp_clause_code_name[] =
@@ -413,7 +415,9 @@ const char * const omp_clause_code_name[] =
   "num_workers",
   "vector_length",
   "tile",
-  "_griddim_"
+  "_griddim_",
+  "if_present",
+  "finalize",
 };
 
 
@@ -11594,6 +11598,8 @@ walk_tree_1 (tree *tp, walk_tree_fn func, void *data,
 	case OMP_CLAUSE_SEQ:
 	case OMP_CLAUSE_TILE:
 	case OMP_CLAUSE__SIMT_:
+	case OMP_CLAUSE_IF_PRESENT:
+	case OMP_CLAUSE_FINALIZE:
 	  WALK_SUBTREE_TAIL (OMP_CLAUSE_CHAIN (*tp));
 
 	case OMP_CLAUSE_LASTPRIVATE: