diff --git a/libgomp/oacc-mem.c b/libgomp/oacc-mem.c
index 1816b06bf2d..cf054f14b12 100644
--- a/libgomp/oacc-mem.c
+++ b/libgomp/oacc-mem.c
@@ -865,6 +865,23 @@ update_dev_host (int is_dev, void *h, size_t s, int async)
       gomp_fatal ("[%p,%d] is not mapped", h, (int)s);
     }
 
+  if (n->aux && n->aux->attach_count)
+    {
+      size_t nptrs = (n->host_end - n->host_start + sizeof (void *) - 1)
+		     / sizeof (void *);
+      for (size_t i = 0; i < nptrs; i++)
+	if (n->aux->attach_count[i] > 0)
+	  {
+	    gomp_mutex_unlock (&acc_dev->lock);
+	    if (is_dev)
+	      gomp_fatal ("[%p,+%d] device update would overwrite attached "
+			  "pointers", h, (int) s);
+	    else
+	      gomp_fatal ("host update from block [%p,+%d] with attached "
+			  "pointers", h, (int) s);
+	  }
+    }
+
   d = (void *) (n->tgt->tgt_start + n->tgt_offset
 		+ (uintptr_t) h - n->host_start);
 
@@ -1329,11 +1346,26 @@ goacc_exit_data_internal (struct gomp_device_descr *acc_dev, size_t mapnum,
 	    if (copyfrom
 		&& n->refcount != REFCOUNT_INFINITY
 		&& (kind != GOMP_MAP_FROM || n->refcount == 0))
-	      gomp_copy_dev2host (acc_dev, aq, (void *) cur_node.host_start,
-				  (void *) (n->tgt->tgt_start + n->tgt_offset
-					    + cur_node.host_start
-					    - n->host_start),
-				  cur_node.host_end - cur_node.host_start);
+	      {
+		if (n->aux && n->aux->attach_count)
+		  {
+		    size_t nptrs = (n->host_end - n->host_start
+				    + sizeof (void *) - 1) / sizeof (void *);
+		    for (size_t j = 0; j < nptrs; j++)
+		      if (n->aux->attach_count[j] > 0)
+			{
+			  gomp_mutex_unlock (&acc_dev->lock);
+			  gomp_fatal ("copyout of block [%p,+%d] with "
+				      "attached pointers", hostaddrs[i],
+				      (int) size);
+			}
+		  }
+		gomp_copy_dev2host (acc_dev, aq, (void *) cur_node.host_start,
+				    (void *) (n->tgt->tgt_start + n->tgt_offset
+					      + cur_node.host_start
+					      - n->host_start),
+				    cur_node.host_end - cur_node.host_start);
+	      }
 
 	    if (n->refcount == 0)
 	      {
diff --git a/libgomp/target.c b/libgomp/target.c
index badc254a777..db6f56a8ff8 100644
--- a/libgomp/target.c
+++ b/libgomp/target.c
@@ -1649,11 +1649,28 @@ gomp_unmap_vars_internal (struct target_mem_desc *tgt, bool do_copyfrom,
 
       if ((do_unmap && do_copyfrom && tgt->list[i].copy_from)
 	  || tgt->list[i].always_copy_from)
-	gomp_copy_dev2host (devicep, aq,
-			    (void *) (k->host_start + tgt->list[i].offset),
-			    (void *) (k->tgt->tgt_start + k->tgt_offset
-				      + tgt->list[i].offset),
-			    tgt->list[i].length);
+	{
+	  if (k->aux && k->aux->attach_count)
+	    {
+	      size_t nptrs = (k->host_end - k->host_start
+			      + sizeof (void *) - 1) / sizeof (void *);
+	      for (size_t j = 0; j < nptrs; j++)
+		if (k->aux->attach_count[j] > 0)
+		  {
+		    gomp_mutex_unlock (&devicep->lock);
+		    gomp_fatal ("copyout of block [%p,+%d] with "
+				"attached pointers",
+				(void *) (k->host_start + tgt->list[i].offset),
+				(int) (k->host_end - k->host_start));
+		  }
+	    }
+	  gomp_copy_dev2host (devicep, aq,
+			      (void *) (k->host_start + tgt->list[i].offset),
+			      (void *) (k->tgt->tgt_start + k->tgt_offset
+					+ tgt->list[i].offset),
+			      tgt->list[i].length);
+	}
+
       if (do_unmap)
 	{
 	  struct target_mem_desc *k_tgt = k->tgt;
diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/copyback-attached-dynamic-1.c b/libgomp/testsuite/libgomp.oacc-c-c++-common/copyback-attached-dynamic-1.c
new file mode 100644
index 00000000000..bc4e297fa6f
--- /dev/null
+++ b/libgomp/testsuite/libgomp.oacc-c-c++-common/copyback-attached-dynamic-1.c
@@ -0,0 +1,31 @@
+/* { dg-skip-if "" { *-*-* } { "-DACC_MEM_SHARED=1" } } */
+
+#include <stdlib.h>
+
+struct mystruct {
+  int *arr;
+};
+
+int
+main (int argc, char *argv[])
+{
+  int localarray[1024];
+  struct mystruct s;
+  s.arr = localarray;
+
+  #pragma acc enter data copyin(s)
+
+  #pragma acc data copy(s.arr[0:1024])
+  {
+    /* This directive does one too many attachments: it should fail when we try
+       to do the copyout below.  */
+    #pragma acc enter data attach(s.arr)
+    /* { dg-output "copyout of block \\\[0x\[0-9a-f\]+,\\+\[0-9\]+\\\] with attached pointers" } */
+  }
+
+  #pragma acc exit data copyout(s)
+
+  return 0;
+}
+
+/* { dg-shouldfail "" } */
diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/copyback-attached-structural-1.c b/libgomp/testsuite/libgomp.oacc-c-c++-common/copyback-attached-structural-1.c
new file mode 100644
index 00000000000..7846c8c717c
--- /dev/null
+++ b/libgomp/testsuite/libgomp.oacc-c-c++-common/copyback-attached-structural-1.c
@@ -0,0 +1,30 @@
+/* { dg-skip-if "" { *-*-* } { "-DACC_MEM_SHARED=1" } } */
+
+#include <stdlib.h>
+
+struct mystruct {
+  int *arr;
+};
+
+int
+main (int argc, char *argv[])
+{
+  int localarray[1024];
+  struct mystruct s;
+  s.arr = localarray;
+
+  #pragma acc data copy(s)
+  {
+    #pragma acc data copy(s.arr[0:1024])
+    {
+      /* This directive does one too many attachments: it should fail when we try
+	 to do the copyout below.  */
+      #pragma acc enter data attach(s.arr)
+      /* { dg-output "copyout of block \\\[0x\[0-9a-f\]+,\\+\[0-9\]+\\\] with attached pointers" } */
+    }
+  }
+
+  return 0;
+}
+
+/* { dg-shouldfail "" } */
diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/copyback-attached-structural-2.c b/libgomp/testsuite/libgomp.oacc-c-c++-common/copyback-attached-structural-2.c
new file mode 100644
index 00000000000..bffa06eb725
--- /dev/null
+++ b/libgomp/testsuite/libgomp.oacc-c-c++-common/copyback-attached-structural-2.c
@@ -0,0 +1,31 @@
+/* { dg-skip-if "" { *-*-* } { "-DACC_MEM_SHARED=1" } } */
+
+#include <stdlib.h>
+
+struct mystruct {
+  int *arr;
+};
+
+int
+main (int argc, char *argv[])
+{
+  int localarray[1024];
+  struct mystruct s;
+  s.arr = localarray;
+
+  #pragma acc enter data copyin(localarray[0:1024])
+
+  #pragma acc data copy(s)
+  {
+    /* This directive does one too many attachments: it should fail when we try
+       to do the copyout below.  */
+    #pragma acc enter data attach(s.arr)
+    /* { dg-output "copyout of block \\\[0x\[0-9a-f\]+,\\+\[0-9\]+\\\] with attached pointers" } */
+  }
+
+  #pragma acc exit data delete(localarray[0:1024])
+
+  return 0;
+}
+
+/* { dg-shouldfail "" } */
diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/copyback-attached-structural-3.c b/libgomp/testsuite/libgomp.oacc-c-c++-common/copyback-attached-structural-3.c
new file mode 100644
index 00000000000..4b21677af09
--- /dev/null
+++ b/libgomp/testsuite/libgomp.oacc-c-c++-common/copyback-attached-structural-3.c
@@ -0,0 +1,26 @@
+#include <stdlib.h>
+
+struct mystruct {
+  int *arr;
+};
+
+int
+main (int argc, char *argv[])
+{
+  int localarray[1024];
+  struct mystruct s;
+  s.arr = localarray;
+
+  #pragma acc enter data copyin(localarray[0:1024])
+
+  #pragma acc data copy(s)
+  {
+    /* Here the attach and detach balance: this should work.  */
+    #pragma acc enter data attach(s.arr)
+    #pragma acc exit data detach(s.arr)
+  }
+
+  #pragma acc exit data delete(localarray[0:1024])
+
+  return 0;
+}
diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/delete-attached-dynamic-1.c b/libgomp/testsuite/libgomp.oacc-c-c++-common/delete-attached-dynamic-1.c
new file mode 100644
index 00000000000..e074d507fb2
--- /dev/null
+++ b/libgomp/testsuite/libgomp.oacc-c-c++-common/delete-attached-dynamic-1.c
@@ -0,0 +1,26 @@
+#include <stdlib.h>
+
+struct mystruct {
+  int *arr;
+};
+
+int
+main (int argc, char *argv[])
+{
+  int localarray[1024];
+  struct mystruct s;
+  s.arr = localarray;
+
+  #pragma acc enter data copyin(s)
+
+  #pragma acc data copy(s.arr[0:1024])
+  {
+    /* We delete 's' from the target below: this extra attachment is not
+       dangerous and we do not raise an error.  */
+    #pragma acc enter data attach(s.arr)
+  }
+
+  #pragma acc exit data delete(s)
+
+  return 0;
+}
diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/delete-attached-structural-1.c b/libgomp/testsuite/libgomp.oacc-c-c++-common/delete-attached-structural-1.c
new file mode 100644
index 00000000000..e675762ecd8
--- /dev/null
+++ b/libgomp/testsuite/libgomp.oacc-c-c++-common/delete-attached-structural-1.c
@@ -0,0 +1,25 @@
+#include <stdlib.h>
+
+struct mystruct {
+  int *arr;
+};
+
+int
+main (int argc, char *argv[])
+{
+  int localarray[1024];
+  struct mystruct s;
+  s.arr = localarray;
+
+  #pragma acc data copyin(s)
+  {
+    #pragma acc data copy(s.arr[0:1024])
+    {
+      /* This directive does one too many attachments: it should fail when we try
+	 to do the copyout below.  */
+      #pragma acc enter data attach(s.arr)
+    }
+  }
+
+  return 0;
+}
diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/delete-attached-structural-2.c b/libgomp/testsuite/libgomp.oacc-c-c++-common/delete-attached-structural-2.c
new file mode 100644
index 00000000000..d2095255ad3
--- /dev/null
+++ b/libgomp/testsuite/libgomp.oacc-c-c++-common/delete-attached-structural-2.c
@@ -0,0 +1,26 @@
+#include <stdlib.h>
+
+struct mystruct {
+  int *arr;
+};
+
+int
+main (int argc, char *argv[])
+{
+  int localarray[1024];
+  struct mystruct s;
+  s.arr = localarray;
+
+  #pragma acc enter data copyin(localarray[0:1024])
+
+  #pragma acc data copyin(s)
+  {
+    /* We only try to copy in: the extra attachment we're left over with is not
+       harmful and we don't raise an error.  */
+    #pragma acc enter data attach(s.arr)
+  }
+
+  #pragma acc exit data delete(localarray[0:1024])
+
+  return 0;
+}
diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/update-attached-1.c b/libgomp/testsuite/libgomp.oacc-c-c++-common/update-attached-1.c
new file mode 100644
index 00000000000..9f60bfa56f4
--- /dev/null
+++ b/libgomp/testsuite/libgomp.oacc-c-c++-common/update-attached-1.c
@@ -0,0 +1,33 @@
+/* { dg-skip-if "" { *-*-* } { "-DACC_MEM_SHARED=1" } } */
+
+#include <stdlib.h>
+
+struct mystruct {
+  int *arr;
+};
+
+int
+main (int argc, char *argv[])
+{
+  int localarray[1024];
+  int localarray2[1024];
+  struct mystruct s;
+  s.arr = localarray;
+
+  #pragma acc enter data copyin(s)
+
+  #pragma acc data copy(s.arr[0:1024])
+  {
+    s.arr = localarray2;
+    /* This update is dangerous because we have attached pointers: raise an
+       error.  */
+    #pragma acc update device(s)
+    /* { dg-output "\\\[0x\[0-9a-f\]+,\\+\[0-9\]+\\\] device update would overwrite attached pointers" } */
+  }
+
+  #pragma acc exit data delete(s)
+
+  return 0;
+}
+
+/* { dg-shouldfail "" } */
diff --git a/libgomp/testsuite/libgomp.oacc-fortran/deep-copy-6-no_finalize.F90 b/libgomp/testsuite/libgomp.oacc-fortran/deep-copy-6-no_finalize.F90
index ad8da71d7c9..355a381b625 100644
--- a/libgomp/testsuite/libgomp.oacc-fortran/deep-copy-6-no_finalize.F90
+++ b/libgomp/testsuite/libgomp.oacc-fortran/deep-copy-6-no_finalize.F90
@@ -8,7 +8,7 @@
 ! { dg-output ".*CheCKpOInT1(\n|\r\n|\r)" }
 ! { dg-output ".*CheCKpOInT2(\n|\r\n|\r)" }
 
-! Without the finalize, we do not detach properly so the host sees a device
-! pointer, and fails with this STOP code.
-! { dg-output "STOP 7(\n|\r\n|\r)+" }
+! Without the finalize, we do not detach properly and raise an error on attempting
+! the copyout.
+! { dg-output ".*copyout of block \\\[0x\[0-9a-f\]+,\\+\[0-9\]+\\\] with attached pointers(\n|\r\n|\r)+" }
 ! { dg-shouldfail "" }
