From 125621f569cfac9f4caa6afc1976d42b3d21359e Mon Sep 17 00:00:00 2001
From: Thomas Schwinge <thomas@codesourcery.com>
Date: Fri, 29 May 2020 15:22:42 +0200
Subject: [PATCH] [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>
(cherry picked from commit 06ec61726d192659cd446e59a91e78745037f0fd)
---
libgomp/oacc-mem.c | 60 +++++++++++++++++--
.../struct-refcount-1.c | 47 +++++++++++++++
.../libgomp.oacc-fortran/deep-copy-6.f90 | 15 ++---
.../mdc-refcount-1-1-1.f90 | 9 +--
.../mdc-refcount-1-2-1.f90 | 9 +--
.../mdc-refcount-1-2-2.f90 | 9 +--
.../mdc-refcount-1-3-1.f90 | 9 +--
.../mdc-refcount-1-4-1.f90 | 7 ++-
8 files changed, 135 insertions(+), 30 deletions(-)
create mode 100644 libgomp/testsuite/libgomp.oacc-c-c++-common/struct-refcount-1.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;
new file mode 100644
@@ -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;
+}
@@ -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
@@ -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
@@ -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
@@ -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
@@ -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
@@ -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
--
2.26.2