procyberian/gcc
1
0
Fork 0
Code Issues Pull requests Projects Releases Packages Wiki Activity Actions

[OpenACC] Repair/restore 'is_tgt_unmapped' checking

Browse source 
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>
This commit is contained in:
Thomas Schwinge 2020-05-29 15:22:42 +02:00
parent 2112d3242f
commit 06ec61726d
8 changed files with 135 additions and 30 deletions
Show stats Download patch file Download diff file Expand all files Collapse all files

60
libgomp/oacc-mem.c
View file

@ -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;

47
libgomp/testsuite/libgomp.oacc-c-c++-common/struct-refcount-1.c Normal file
View file

@ -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;
}

15
libgomp/testsuite/libgomp.oacc-fortran/deep-copy-6.f90
View file

@ -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

9
libgomp/testsuite/libgomp.oacc-fortran/mdc-refcount-1-1-1.f90
View file

@ -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

9
libgomp/testsuite/libgomp.oacc-fortran/mdc-refcount-1-2-1.f90
View file

@ -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

9
libgomp/testsuite/libgomp.oacc-fortran/mdc-refcount-1-2-2.f90
View file

@ -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

9
libgomp/testsuite/libgomp.oacc-fortran/mdc-refcount-1-3-1.f90
View file

@ -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

7
libgomp/testsuite/libgomp.oacc-fortran/mdc-refcount-1-4-1.f90
View file

@ -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