From: Julian Brown <julian@codesourcery.com>
Date: Thu, 2 Jul 2020 21:18:20 +0000 (-0700)
Subject: openacc: Remove unnecessary detach finalization
X-Git-Url: https://git.libre-soc.org/?a=commitdiff_plain;h=25bce75c77ec5617c78173d837d3b664c0f20968;p=gcc.git

openacc: Remove unnecessary detach finalization

The call to gomp_detach_pointer in gomp_unmap_vars_internal does not
need to force finalization, and doing so may mask mismatched pointer
attachments/detachments. This patch removes the forcing.

2020-07-16  Julian Brown  <julian@codesourcery.com>
	    Thomas Schwinge  <thomas@codesourcery.com>

libgomp/
	* target.c (gomp_unmap_vars_internal): Remove unnecessary forcing of
	finalization for detach operation.
	* testsuite/libgomp.oacc-c-c++-common/structured-detach-underflow.c:
	New test.

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

diff --git a/libgomp/target.c b/libgomp/target.c
index d6b3572c8d8..00c75fbd885 100644
--- a/libgomp/target.c
+++ b/libgomp/target.c
@@ -1437,7 +1437,7 @@ gomp_unmap_vars_internal (struct target_mem_desc *tgt, bool do_copyfrom,
       if (k != NULL && tgt->list[i].do_detach)
 	gomp_detach_pointer (devicep, aq, k, tgt->list[i].key->host_start
 					     + tgt->list[i].offset,
-			     k->refcount == 1, NULL);
+			     false, NULL);
     }
 
   for (i = 0; i < tgt->list_count; i++)
diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/structured-detach-underflow.c b/libgomp/testsuite/libgomp.oacc-c-c++-common/structured-detach-underflow.c
new file mode 100644
index 00000000000..fc1f59e2185
--- /dev/null
+++ b/libgomp/testsuite/libgomp.oacc-c-c++-common/structured-detach-underflow.c
@@ -0,0 +1,28 @@
+/* { dg-skip-if "" { *-*-* } { "*" } { "-DACC_MEM_SHARED=0" } } */
+
+#include <stdio.h>
+
+int main ()
+{
+  struct {
+    int *arr;
+  } mystr;
+  int localarr[16];
+  mystr.arr = localarr;
+
+  #pragma acc enter data copyin(mystr, localarr[0:16])
+
+  #pragma acc data attach(mystr.arr)
+  {
+    #pragma acc exit data detach(mystr.arr)
+    fprintf (stderr, "CheCKpOInT1\n");
+    /* { dg-output ".*CheCKpOInT1(\n|\r\n|\r)" } */
+  }
+  /* { dg-shouldfail "" }
+     { dg-output "(\n|\r\n|\r)libgomp: attach count underflow(\n|\r\n|\r)$" } */
+  fprintf (stderr, "CheCKpOInT2\n");
+
+  #pragma acc exit data copyout(mystr, localarr[0:16])
+
+  return 0;
+}