diff options
| author | Julian Brown <julian@codesourcery.com> | 2020-07-02 14:18:20 -0700 |
|---|---|---|
| committer | Julian Brown <julian@codesourcery.com> | 2020-07-23 12:50:07 -0700 |
| commit | 25bce75c77ec5617c78173d837d3b664c0f20968 (patch) | |
| tree | 1c4a8416c0477d3e89e6b8d0aee7ac277d025ee8 | |
| parent | d4b3ea941b1a15cadb79e7ff8a36891131f8804b (diff) | |
| download | gcc-25bce75c77ec5617c78173d837d3b664c0f20968.zip gcc-25bce75c77ec5617c78173d837d3b664c0f20968.tar.gz gcc-25bce75c77ec5617c78173d837d3b664c0f20968.tar.bz2 | |
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>
| -rw-r--r-- | libgomp/target.c | 2 | ||||
| -rw-r--r-- | libgomp/testsuite/libgomp.oacc-c-c++-common/structured-detach-underflow.c | 28 |
2 files changed, 29 insertions, 1 deletions
diff --git a/libgomp/target.c b/libgomp/target.c index d6b3572..00c75fb 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 0000000..fc1f59e --- /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; +} |