[OpenACC] Repair/restore 'is_tgt_unmapped' checking

	libgomp/
	* oacc-mem.c (goacc_exit_datum): Repair 'is_tgt_unmapped'
	checking.
	(acc_unmap_data, goacc_exit_data_internal): Restore
	'is_tgt_unmapped' checking.
	* testsuite/libgomp.oacc-c-c++-common/struct-refcount-1.c: New
	file.
	* testsuite/libgomp.oacc-fortran/deep-copy-6.f90: Adjust.
	* testsuite/libgomp.oacc-fortran/mdc-refcount-1-1-1.f90: Likewise.
	* testsuite/libgomp.oacc-fortran/mdc-refcount-1-2-1.f90: Likewise.
	* testsuite/libgomp.oacc-fortran/mdc-refcount-1-2-2.f90: Likewise.
	* testsuite/libgomp.oacc-fortran/mdc-refcount-1-3-1.f90: Likewise.
	* testsuite/libgomp.oacc-fortran/mdc-refcount-1-4-1.f90: Likewise.

Co-Authored-By: Julian Brown <julian@codesourcery.com>

diff --git a/libgomp/oacc-mem.c b/libgomp/oacc-mem.c
index 8e8c7c3093d56380656e3d45bcd893119bfa8141..b7c85cf5976f2d2de5aa145bbd476416810a2c7e 100644 (file)
--- a/libgomp/oacc-mem.c
+++ b/libgomp/oacc-mem.c
@@ -485,7 +485,8 @@ acc_unmap_data (void *h)
   tgt->tgt_end = 0;
   tgt->to_free = NULL;
 
-  gomp_remove_var (acc_dev, n);
+  bool is_tgt_unmapped = gomp_remove_var (acc_dev, n);
+  assert (is_tgt_unmapped);
 
   gomp_mutex_unlock (&acc_dev->lock);
 
@@ -727,8 +728,16 @@ goacc_exit_datum (void *h, size_t s, unsigned short kind, int async)
        gomp_remove_var_async (acc_dev, n, aq);
       else
        {
+         size_t num_mappings = 0;
+         /* If the target_mem_desc represents a single data mapping, we can
+            check that it is freed when this splay tree key's refcount reaches
+            zero.  Otherwise (e.g. for a 'GOMP_MAP_STRUCT' mapping with
+            multiple members), fall back to skipping the test.  */
+         for (size_t l_i = 0; l_i < n->tgt->list_count; ++l_i)
+           if (n->tgt->list[l_i].key)
+             ++num_mappings;
          bool is_tgt_unmapped = gomp_remove_var (acc_dev, n);
-         assert (is_tgt_unmapped);
+         assert (is_tgt_unmapped || num_mappings > 1);
        }
     }
 
@@ -1145,7 +1154,28 @@ goacc_exit_data_internal (struct gomp_device_descr *acc_dev, size_t mapnum,
                                  cur_node.host_end - cur_node.host_start);
 
            if (n->refcount == 0)
-             gomp_remove_var_async (acc_dev, n, aq);
+             {
+               if (aq)
+                 /* TODO We can't do the 'is_tgt_unmapped' checking -- see the
+                    'gomp_unref_tgt' comment in
+                    <http://mid.mail-archive.com/878snl36eu.fsf@euler.schwinge.homeip.net>;
+                    PR92881.  */
+                 gomp_remove_var_async (acc_dev, n, aq);
+               else
+                 {
+                   size_t num_mappings = 0;
+                   /* If the target_mem_desc represents a single data mapping,
+                      we can check that it is freed when this splay tree key's
+                      refcount reaches zero.  Otherwise (e.g. for a
+                      'GOMP_MAP_STRUCT' mapping with multiple members), fall
+                      back to skipping the test.  */
+                   for (size_t l_i = 0; l_i < n->tgt->list_count; ++l_i)
+                     if (n->tgt->list[l_i].key)
+                       ++num_mappings;
+                   bool is_tgt_unmapped = gomp_remove_var (acc_dev, n);
+                   assert (is_tgt_unmapped || num_mappings > 1);
+                 }
+             }
          }
          break;
 
@@ -1177,7 +1207,29 @@ goacc_exit_data_internal (struct gomp_device_descr *acc_dev, size_t mapnum,
                             && str->refcount != REFCOUNT_INFINITY)
                      str->refcount--;
                    if (str->refcount == 0)
-                     gomp_remove_var_async (acc_dev, str, aq);
+                     {
+                       if (aq)
+                         /* TODO We can't do the 'is_tgt_unmapped' checking --
+                            see the 'gomp_unref_tgt' comment in
+                            <http://mid.mail-archive.com/878snl36eu.fsf@euler.schwinge.homeip.net>;
+                            PR92881.  */
+                         gomp_remove_var_async (acc_dev, str, aq);
+                       else
+                         {
+                           size_t num_mappings = 0;
+                           /* If the target_mem_desc represents a single data
+                              mapping, we can check that it is freed when this
+                              splay tree key's refcount reaches zero.
+                              Otherwise (e.g. for a 'GOMP_MAP_STRUCT' mapping
+                              with multiple members), fall back to skipping
+                              the test.  */
+                           for (size_t l_i = 0; l_i < str->tgt->list_count; ++l_i)
+                             if (str->tgt->list[l_i].key)
+                               ++num_mappings;
+                           bool is_tgt_unmapped = gomp_remove_var (acc_dev, str);
+                           assert (is_tgt_unmapped || num_mappings > 1);
+                         }
+                     }
                  }
              }
            i += elems;

diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/struct-refcount-1.c b/libgomp/testsuite/libgomp.oacc-c-c++-common/struct-refcount-1.c
new file mode 100644 (file)
index 0000000..bde5890
--- /dev/null
+++ b/libgomp/testsuite/libgomp.oacc-c-c++-common/struct-refcount-1.c
@@ -0,0 +1,47 @@
+/* Test dynamic unmapping of separate structure members.  */
+
+/* { dg-skip-if "" { *-*-* } { "-DACC_MEM_SHARED=1" } } */
+
+#include <assert.h>
+#include <stdbool.h>
+#include <openacc.h>
+
+struct s
+{
+  char a;
+  float b;
+};
+
+void test (bool use_directives)
+{
+  struct s s;
+
+#pragma acc enter data create(s.a, s.b)
+  assert (acc_is_present (&s.a, sizeof s.a));
+  assert (acc_is_present (&s.b, sizeof s.b));
+
+  if (use_directives)
+    {
+#pragma acc exit data delete(s.a)
+    }
+  else
+    acc_delete (&s.a, sizeof s.a);
+  assert (!acc_is_present (&s.a, sizeof s.a));
+  assert (acc_is_present (&s.b, sizeof s.b));
+  if (use_directives)
+    {
+#pragma acc exit data delete(s.b)
+    }
+  else
+    acc_delete (&s.b, sizeof s.b);
+  assert (!acc_is_present (&s.a, sizeof s.a));
+  assert (!acc_is_present (&s.b, sizeof s.b));
+}
+
+int main ()
+{
+  test (true);
+  test (false);
+
+  return 0;
+}

diff --git a/libgomp/testsuite/libgomp.oacc-fortran/deep-copy-6.f90 b/libgomp/testsuite/libgomp.oacc-fortran/deep-copy-6.f90
index a7943d93d542f3e56a046676e99f3af1195930f1..5837a4039106045ed43dbf1c26ec09f94e03b29c 100644 (file)
--- a/libgomp/testsuite/libgomp.oacc-fortran/deep-copy-6.f90
+++ b/libgomp/testsuite/libgomp.oacc-fortran/deep-copy-6.f90
@@ -40,19 +40,20 @@ program dtype
   if (.not. acc_is_present(var%a(5:n - 5))) stop 11
   if (.not. acc_is_present(var%b(5:n - 5))) stop 12
   if (.not. acc_is_present(var)) stop 13
-!$acc exit data copyout(var%a(5:n - 5), var%b(5:n - 5)) finalize
-  if (acc_get_device_type() .ne. acc_device_host) then
-     if (acc_is_present(var%a(5:n - 5))) stop 21
-     if (acc_is_present(var%b(5:n - 5))) stop 22
-  end if
   print *, "CheCKpOInT1"
   ! { dg-output ".*CheCKpOInT1(\n|\r\n|\r)" }
-  if (.not. acc_is_present(var)) stop 23
-  !TODO { dg-output "STOP 23(\n|\r\n|\r)$" { target { ! openacc_host_selected } } } ! Scan for what we expect in the "XFAILed" case (without actually XFAILing).
+!$acc exit data copyout(var%a(5:n - 5), var%b(5:n - 5)) finalize
+  !TODO     goacc_exit_data_internal: Assertion `is_tgt_unmapped || num_mappings > 1' failed.
+  !TODO { dg-output ".*\[Aa\]ssert.*is_tgt_unmapped" { target { ! openacc_host_selected } } } ! Scan for what we expect in the "XFAILed" case (without actually XFAILing).
   !TODO { dg-shouldfail "XFAILed" { ! openacc_host_selected } } ! ... instead of 'dg-xfail-run-if' so that 'dg-output' is evaluated at all.
   !TODO { dg-final { if { [dg-process-target { xfail { ! openacc_host_selected } }] == "F" } { xfail "[testname-for-summary] really is XFAILed" } } } ! ... so that we still get an XFAIL visible in the log.
   print *, "CheCKpOInT2"
   ! { dg-output ".CheCKpOInT2(\n|\r\n|\r)" { target { openacc_host_selected } } }
+  if (acc_get_device_type() .ne. acc_device_host) then
+     if (acc_is_present(var%a(5:n - 5))) stop 21
+     if (acc_is_present(var%b(5:n - 5))) stop 22
+  end if
+  if (.not. acc_is_present(var)) stop 23
 
 !$acc end data
 

diff --git a/libgomp/testsuite/libgomp.oacc-fortran/mdc-refcount-1-1-1.f90 b/libgomp/testsuite/libgomp.oacc-fortran/mdc-refcount-1-1-1.f90
index 449a6cf6894e70c8e19e34cc68eba7f2930721d9..445cbabb8ca4a0ddfdbebb6638ea86341561d631 100644 (file)
--- a/libgomp/testsuite/libgomp.oacc-fortran/mdc-refcount-1-1-1.f90
+++ b/libgomp/testsuite/libgomp.oacc-fortran/mdc-refcount-1-1-1.f90
@@ -21,16 +21,17 @@ program main
   if (.not. acc_is_present(var%a)) stop 1
   if (.not. acc_is_present(var)) stop 2
 
-  !$acc exit data delete(var%a) finalize
-  if (acc_is_present(var%a)) stop 3
   print *, "CheCKpOInT1"
   ! { dg-output ".*CheCKpOInT1(\n|\r\n|\r)" }
-  if (.not. acc_is_present(var)) stop 4
-  !TODO { dg-output "STOP 4(\n|\r\n|\r)$" { target { ! openacc_host_selected } } } ! Scan for what we expect in the "XFAILed" case (without actually XFAILing).
+  !$acc exit data delete(var%a) finalize
+  !TODO     goacc_exit_data_internal: Assertion `is_tgt_unmapped || num_mappings > 1' failed.
+  !TODO { dg-output ".*\[Aa\]ssert.*is_tgt_unmapped" { target { ! openacc_host_selected } } } ! Scan for what we expect in the "XFAILed" case (without actually XFAILing).
   !TODO { dg-shouldfail "XFAILed" { ! openacc_host_selected } } ! ... instead of 'dg-xfail-run-if' so that 'dg-output' is evaluated at all.
   !TODO { dg-final { if { [dg-process-target { xfail { ! openacc_host_selected } }] == "F" } { xfail "[testname-for-summary] really is XFAILed" } } } ! ... so that we still get an XFAIL visible in the log.
   print *, "CheCKpOInT2"
   ! { dg-output ".CheCKpOInT2(\n|\r\n|\r)" { target { openacc_host_selected } } }
+  if (acc_is_present(var%a)) stop 3
+  if (.not. acc_is_present(var)) stop 4
 
   !$acc end data
   if (acc_is_present(var%a)) stop 5

diff --git a/libgomp/testsuite/libgomp.oacc-fortran/mdc-refcount-1-2-1.f90 b/libgomp/testsuite/libgomp.oacc-fortran/mdc-refcount-1-2-1.f90
index a7e649d3041529e9240eaada554685eb005a7dad..8554534b2f271cc48565f9beeaf48238bcf1fb38 100644 (file)
--- a/libgomp/testsuite/libgomp.oacc-fortran/mdc-refcount-1-2-1.f90
+++ b/libgomp/testsuite/libgomp.oacc-fortran/mdc-refcount-1-2-1.f90
@@ -23,16 +23,17 @@ program main
   if (.not. acc_is_present(var%a)) stop 1
   if (.not. acc_is_present(var)) stop 2
 
-  !$acc exit data delete(var%a) finalize
-  if (acc_is_present(var%a)) stop 3
   print *, "CheCKpOInT1"
   ! { dg-output ".*CheCKpOInT1(\n|\r\n|\r)" }
-  if (.not. acc_is_present(var)) stop 4
-  !TODO { dg-output "STOP 4(\n|\r\n|\r)$" { target { ! openacc_host_selected } } } ! Scan for what we expect in the "XFAILed" case (without actually XFAILing).
+  !$acc exit data delete(var%a) finalize
+  !TODO     goacc_exit_data_internal: Assertion `is_tgt_unmapped || num_mappings > 1' failed.
+  !TODO { dg-output ".*\[Aa\]ssert.*is_tgt_unmapped" { target { ! openacc_host_selected } } } ! Scan for what we expect in the "XFAILed" case (without actually XFAILing).
   !TODO { dg-shouldfail "XFAILed" { ! openacc_host_selected } } ! ... instead of 'dg-xfail-run-if' so that 'dg-output' is evaluated at all.
   !TODO { dg-final { if { [dg-process-target { xfail { ! openacc_host_selected } }] == "F" } { xfail "[testname-for-summary] really is XFAILed" } } } ! ... so that we still get an XFAIL visible in the log.
   print *, "CheCKpOInT2"
   ! { dg-output ".CheCKpOInT2(\n|\r\n|\r)" { target { openacc_host_selected } } }
+  if (acc_is_present(var%a)) stop 3
+  if (.not. acc_is_present(var)) stop 4
 
   !$acc end data
   if (acc_is_present(var%a)) stop 5

diff --git a/libgomp/testsuite/libgomp.oacc-fortran/mdc-refcount-1-2-2.f90 b/libgomp/testsuite/libgomp.oacc-fortran/mdc-refcount-1-2-2.f90
index 3402fafc7e28da136fe9a75064b265ab97fac439..8e696cc70e80881ca94b6edf77b4ce2ee6103a9d 100644 (file)
--- a/libgomp/testsuite/libgomp.oacc-fortran/mdc-refcount-1-2-2.f90
+++ b/libgomp/testsuite/libgomp.oacc-fortran/mdc-refcount-1-2-2.f90
@@ -23,16 +23,17 @@ program main
   if (.not. acc_is_present(var%a)) stop 1
   if (.not. acc_is_present(var)) stop 2
 
-  !$acc exit data delete(var%a)
-  if (acc_is_present(var%a)) stop 3
   print *, "CheCKpOInT1"
   ! { dg-output ".*CheCKpOInT1(\n|\r\n|\r)" }
-  if (.not. acc_is_present(var)) stop 4
-  !TODO { dg-output "STOP 4(\n|\r\n|\r)$" { target { ! openacc_host_selected } } } ! Scan for what we expect in the "XFAILed" case (without actually XFAILing).
+  !$acc exit data delete(var%a)
+  !TODO     goacc_exit_data_internal: Assertion `is_tgt_unmapped || num_mappings > 1' failed.
+  !TODO { dg-output ".*\[Aa\]ssert.*is_tgt_unmapped" { target { ! openacc_host_selected } } } ! Scan for what we expect in the "XFAILed" case (without actually XFAILing).
   !TODO { dg-shouldfail "XFAILed" { ! openacc_host_selected } } ! ... instead of 'dg-xfail-run-if' so that 'dg-output' is evaluated at all.
   !TODO { dg-final { if { [dg-process-target { xfail { ! openacc_host_selected } }] == "F" } { xfail "[testname-for-summary] really is XFAILed" } } } ! ... so that we still get an XFAIL visible in the log.
   print *, "CheCKpOInT2"
   ! { dg-output ".CheCKpOInT2(\n|\r\n|\r)" { target { openacc_host_selected } } }
+  if (acc_is_present(var%a)) stop 3
+  if (.not. acc_is_present(var)) stop 4
 
   !$acc end data
   if (acc_is_present(var%a)) stop 5

diff --git a/libgomp/testsuite/libgomp.oacc-fortran/mdc-refcount-1-3-1.f90 b/libgomp/testsuite/libgomp.oacc-fortran/mdc-refcount-1-3-1.f90
index 7504969d9a5629b55ce055f7c37267fa7e367c38..070a6f8e1493fbe8772a92c6d858b61502715f21 100644 (file)
--- a/libgomp/testsuite/libgomp.oacc-fortran/mdc-refcount-1-3-1.f90
+++ b/libgomp/testsuite/libgomp.oacc-fortran/mdc-refcount-1-3-1.f90
@@ -24,16 +24,17 @@ program main
   if (.not. acc_is_present(var)) stop 2
 
   !$acc exit data detach(var%a)
-  !$acc exit data delete(var%a) finalize
-  if (acc_is_present(var%a)) stop 3
   print *, "CheCKpOInT1"
   ! { dg-output ".*CheCKpOInT1(\n|\r\n|\r)" }
-  if (.not. acc_is_present(var)) stop 4
-  !TODO { dg-output "STOP 4(\n|\r\n|\r)$" { target { ! openacc_host_selected } } } ! Scan for what we expect in the "XFAILed" case (without actually XFAILing).
+  !$acc exit data delete(var%a) finalize
+  !TODO     goacc_exit_data_internal: Assertion `is_tgt_unmapped || num_mappings > 1' failed.
+  !TODO { dg-output ".*\[Aa\]ssert.*is_tgt_unmapped" { target { ! openacc_host_selected } } } ! Scan for what we expect in the "XFAILed" case (without actually XFAILing).
   !TODO { dg-shouldfail "XFAILed" { ! openacc_host_selected } } ! ... instead of 'dg-xfail-run-if' so that 'dg-output' is evaluated at all.
   !TODO { dg-final { if { [dg-process-target { xfail { ! openacc_host_selected } }] == "F" } { xfail "[testname-for-summary] really is XFAILed" } } } ! ... so that we still get an XFAIL visible in the log.
   print *, "CheCKpOInT2"
   ! { dg-output ".CheCKpOInT2(\n|\r\n|\r)" { target { openacc_host_selected } } }
+  if (acc_is_present(var%a)) stop 3
+  if (.not. acc_is_present(var)) stop 4
 
   !$acc end data
   if (acc_is_present(var%a)) stop 5

diff --git a/libgomp/testsuite/libgomp.oacc-fortran/mdc-refcount-1-4-1.f90 b/libgomp/testsuite/libgomp.oacc-fortran/mdc-refcount-1-4-1.f90
index fedae0db054bcf90a331fe5aba927645d4b0d5f9..b22e411567ffb9a649624e32afb92001a52bf8ae 100644 (file)
--- a/libgomp/testsuite/libgomp.oacc-fortran/mdc-refcount-1-4-1.f90
+++ b/libgomp/testsuite/libgomp.oacc-fortran/mdc-refcount-1-4-1.f90
@@ -23,15 +23,16 @@ program main
   if (.not. acc_is_present(var%a)) stop 1
   if (.not. acc_is_present(var)) stop 2
 
-  !$acc exit data detach(var%a) finalize
   print *, "CheCKpOInT1"
   ! { dg-output ".*CheCKpOInT1(\n|\r\n|\r)" }
-  !$acc exit data delete(var%a)
-  !TODO { dg-output "(\n|\r\n|\r)libgomp: struct not mapped for detach operation(\n|\r\n|\r)$" { target { ! openacc_host_selected } } } ! Scan for what we expect in the "XFAILed" case (without actually XFAILing).
+  !$acc exit data detach(var%a) finalize
+  !TODO     goacc_exit_data_internal: Assertion `is_tgt_unmapped || num_mappings > 1' failed.
+  !TODO { dg-output ".*\[Aa\]ssert.*is_tgt_unmapped" { target { ! openacc_host_selected } } } ! Scan for what we expect in the "XFAILed" case (without actually XFAILing).
   !TODO { dg-shouldfail "XFAILed" { ! openacc_host_selected } } ! ... instead of 'dg-xfail-run-if' so that 'dg-output' is evaluated at all.
   !TODO { dg-final { if { [dg-process-target { xfail { ! openacc_host_selected } }] == "F" } { xfail "[testname-for-summary] really is XFAILed" } } } ! ... so that we still get an XFAIL visible in the log.
   print *, "CheCKpOInT2"
   ! { dg-output ".CheCKpOInT2(\n|\r\n|\r)" { target { openacc_host_selected } } }
+  !$acc exit data delete(var%a)
   if (acc_is_present(var%a)) stop 3
   if (.not. acc_is_present(var)) stop 4