Message ID | 87k08p48mb.fsf@euler.schwinge.homeip.net |
---|---|
State | New |
Headers | show |
Series | Enhance 'libgomp.c-c++-common/requires-4.c', 'libgomp.c-c++-common/requires-5.c' testing (was: [Patch][v4] OpenMP: Move omp requires checks to libgomp) | expand |
On 07.07.22 10:42, Thomas Schwinge wrote: > In preparation for other changes: ... > On 2022-06-29T16:33:02+0200, Tobias Burnus<tobias@codesourcery.com> wrote: >> +/* { dg-output "devices present but 'omp requires unified_address, unified_shared_memory, reverse_offload' cannot be fulfilled" } */ > (The latter diagnostic later got conditionalized by 'GOMP_DEBUG=1'.) > OK to push the attached "Enhance 'libgomp.c-c++-common/requires-4.c', > 'libgomp.c-c++-common/requires-5.c' testing"? ... > libgomp/ > * testsuite/libgomp.c-c++-common/requires-4.c: Enhance testing. > * testsuite/libgomp.c-c++-common/requires-5.c: Likewise. ... > --- a/libgomp/testsuite/libgomp.c-c++-common/requires-4.c > +++ b/libgomp/testsuite/libgomp.c-c++-common/requires-4.c > @@ -1,22 +1,29 @@ > -/* { dg-do link { target offloading_enabled } } */ > /* { dg-additional-options "-flto" } */ > /* { dg-additional-sources requires-4-aux.c } */ > > -/* Check diagnostic by device-compiler's or host compiler's lto1. > +/* Check no diagnostic by device-compiler's or host compiler's lto1. I note that without ENABLE_OFFLOADING that there is never any lto1 diagnostic. However, given that no diagnostic is expected, it also works for "! offloading_enabled". Thus, the change fine. > Other file uses: 'requires reverse_offload', but that's inactive as > there are no declare target directives, device constructs nor device routines */ > > +/* For actual offload execution, prints the following (only) if GOMP_DEBUG=1: > + "devices present but 'omp requires unified_address, unified_shared_memory, reverse_offload' cannot be fulfilled" > + and does host-fallback execution. */ The latter is only true when also device code is produced – and a device is available for that/those device types. I think that's what you imply by "For actual offload execution", but it is a bit hidden. Maybe s/For actual offload execution, prints/It may print/ is clearer? In principle, it would be nice if we could test for the output, but currently setting an env var for remote execution does not work, yet. Cf. https://gcc.gnu.org/pipermail/gcc-patches/2022-July/597773.html - When set, we could use offload_target_nvptx etc. (..._amdgcn, ..._any) to test – as this guarantees that it is compiled for that device + the device is available. > + > #pragma omp requires unified_address,unified_shared_memory > > -int a[10]; > +int a[10] = { 0 }; > extern void foo (void); > > int > main (void) > { > - #pragma omp target > + #pragma omp target map(to: a) Hmm, I wonder whether I like it or not. Without, there is an implicit "map(tofrom:a)". On the other hand, OpenMP permits that – even with unified-shared memory – the implementation my copy the data to the device. (For instance, to permit faster access to "a".) Thus, ... > + for (int i = 0; i < 10; i++) > + a[i] = i; > + > for (int i = 0; i < 10; i++) > - a[i] = 0; > + if (a[i] != i) > + __builtin_abort (); ... this condition (back on the host) could also fail with USM. However, given that to my knowledge no USM implementation actually copies the data, I believe it is fine. (Disclaimer: I have not checked what OG12, but I guess it also does not copy it.) > --- a/libgomp/testsuite/libgomp.c-c++-common/requires-5.c > +++ b/libgomp/testsuite/libgomp.c-c++-common/requires-5.c > @@ -1,21 +1,25 @@ > -/* { dg-do run { target { offload_target_nvptx || offload_target_amdgcn } } } */ > /* { dg-additional-sources requires-5-aux.c } */ > > +/* For actual offload execution, prints the following (only) if GOMP_DEBUG=1: > + "devices present but 'omp requires unified_address, unified_shared_memory, reverse_offload' cannot be fulfilled" > + and does host-fallback execution. */ > + This wording is correct with the now-removed check – but if you remove the offload_target..., it only "might" print it, depending, well, on the conditions set by offload_target... > #pragma omp requires unified_shared_memory, unified_address, reverse_offload > > -int a[10]; > +int a[10] = { 0 }; > extern void foo (void); > > int > main (void) > { > - #pragma omp target > + #pragma omp target map(to: a) > + for (int i = 0; i < 10; i++) > + a[i] = i; > + > for (int i = 0; i < 10; i++) > - a[i] = 0; > + if (a[i] != i) > + __builtin_abort (); > > foo (); > return 0; > } Thus: LGTM, if you update the GOMP_DEBUG=... wording, either using "might" (etc.) or by being more explicit. Once we have remote setenv, we probably want to add another testcase to check for the GOMP_DEBUG=1, copying an existing one, adding dg-output and restricting it to target offload_target_... Tobias ----------------- Siemens Electronic Design Automation GmbH; Anschrift: Arnulfstraße 201, 80634 München; Gesellschaft mit beschränkter Haftung; Geschäftsführer: Thomas Heurung, Frank Thürauf; Sitz der Gesellschaft: München; Registergericht München, HRB 106955
Hi Tobias! On 2022-07-07T11:36:34+0200, Tobias Burnus <tobias@codesourcery.com> wrote: > On 07.07.22 10:42, Thomas Schwinge wrote: >> In preparation for other changes: > ... >> On 2022-06-29T16:33:02+0200, Tobias Burnus<tobias@codesourcery.com> wrote: >>> +/* { dg-output "devices present but 'omp requires unified_address, unified_shared_memory, reverse_offload' cannot be fulfilled" } */ >> (The latter diagnostic later got conditionalized by 'GOMP_DEBUG=1'.) >> OK to push the attached "Enhance 'libgomp.c-c++-common/requires-4.c', >> 'libgomp.c-c++-common/requires-5.c' testing"? > ... >> libgomp/ >> * testsuite/libgomp.c-c++-common/requires-4.c: Enhance testing. >> * testsuite/libgomp.c-c++-common/requires-5.c: Likewise. > ... >> --- a/libgomp/testsuite/libgomp.c-c++-common/requires-4.c >> +++ b/libgomp/testsuite/libgomp.c-c++-common/requires-4.c >> @@ -1,22 +1,29 @@ >> -/* { dg-do link { target offloading_enabled } } */ >> /* { dg-additional-options "-flto" } */ >> /* { dg-additional-sources requires-4-aux.c } */ >> >> -/* Check diagnostic by device-compiler's or host compiler's lto1. >> +/* Check no diagnostic by device-compiler's or host compiler's lto1. > > I note that without ENABLE_OFFLOADING that there is never any lto1 > diagnostic. > > However, given that no diagnostic is expected, it also works for "! > offloading_enabled". > > Thus, the change fine. ACK. >> Other file uses: 'requires reverse_offload', but that's inactive as >> there are no declare target directives, device constructs nor device routines */ >> >> +/* For actual offload execution, prints the following (only) if GOMP_DEBUG=1: >> + "devices present but 'omp requires unified_address, unified_shared_memory, reverse_offload' cannot be fulfilled" >> + and does host-fallback execution. */ > > The latter is only true when also device code is produced – and a device > is available for that/those device types. I think that's what you imply > by "For actual offload execution" ACK. > but it is a bit hidden. > > Maybe s/For actual offload execution, prints/It may print/ is clearer? I've settled on: /* Depending on offload device capabilities, it may print something like the following (only) if GOMP_DEBUG=1: "devices present but 'omp requires unified_address, unified_shared_memory, reverse_offload' cannot be fulfilled" and in that case does host-fallback execution. */ > In principle, it would be nice if we could test for the output, but > currently setting an env var for remote execution does not work, yet. > Cf. https://gcc.gnu.org/pipermail/gcc-patches/2022-July/597773.html Right, I'm aware of that issue with remote testing, and that's why I didn't propose such output verification. (In a few other test cases, we do have 'dg-set-target-env-var GOMP_DEBUG "1"', which then at present are UNSUPPORTED for remote testing.) > When set, we could use offload_target_nvptx etc. (..._amdgcn, ..._any) > to test – as this guarantees that it is compiled for that device + the > device is available. Use 'target offload_device_nvptx', not 'target offload_target_nvptx', etc. ;-) >> + >> #pragma omp requires unified_address,unified_shared_memory >> >> -int a[10]; >> +int a[10] = { 0 }; >> extern void foo (void); >> >> int >> main (void) >> { >> - #pragma omp target >> + #pragma omp target map(to: a) > > Hmm, I wonder whether I like it or not. Without, there is an implicit > "map(tofrom:a)". On the other hand, OpenMP permits that – even with > unified-shared memory – the implementation my copy the data to the > device. (For instance, to permit faster access to "a".) > > Thus, ... > >> + for (int i = 0; i < 10; i++) >> + a[i] = i; >> + >> for (int i = 0; i < 10; i++) >> - a[i] = 0; >> + if (a[i] != i) >> + __builtin_abort (); > ... this condition (back on the host) could also fail with USM. However, > given that to my knowledge no USM implementation actually copies the > data, I believe it is fine. Right, this is meant to describe/test the current GCC master branch behavior, where USM isn't supported, so I didn't consider that. But I agree, a source code comment should be added: As no offload devices support USM at present, we may verify host-fallback execution by absence of separate memory spaces. */ > (Disclaimer: I have not checked what OG12, > but I guess it also does not copy it.) >> --- a/libgomp/testsuite/libgomp.c-c++-common/requires-5.c >> +++ b/libgomp/testsuite/libgomp.c-c++-common/requires-5.c >> @@ -1,21 +1,25 @@ >> -/* { dg-do run { target { offload_target_nvptx || offload_target_amdgcn } } } */ >> /* { dg-additional-sources requires-5-aux.c } */ >> >> +/* For actual offload execution, prints the following (only) if GOMP_DEBUG=1: >> + "devices present but 'omp requires unified_address, unified_shared_memory, reverse_offload' cannot be fulfilled" >> + and does host-fallback execution. */ >> + > This wording is correct with the now-removed check – but if you remove > the offload_target..., it only "might" print it, depending, well, on the > conditions set by offload_target... >> #pragma omp requires unified_shared_memory, unified_address, reverse_offload >> >> -int a[10]; >> +int a[10] = { 0 }; >> extern void foo (void); >> >> int >> main (void) >> { >> - #pragma omp target >> + #pragma omp target map(to: a) >> + for (int i = 0; i < 10; i++) >> + a[i] = i; >> + >> for (int i = 0; i < 10; i++) >> - a[i] = 0; >> + if (a[i] != i) >> + __builtin_abort (); >> >> foo (); >> return 0; >> } > > Thus: LGTM, if you update the GOMP_DEBUG=... wording, either using > "might" (etc.) or by being more explicit. Used the wording as in 'libgomp.c-c++-common/requires-4.c'. Thanks for the review! Pushed to master branch commit 5647e2c3853cbd51a6536a84b8eb0eb3c210dfbf "Enhance 'libgomp.c-c++-common/requires-4.c', 'libgomp.c-c++-common/requires-5.c' testing", see attached. Grüße Thomas ----------------- Siemens Electronic Design Automation GmbH; Anschrift: Arnulfstraße 201, 80634 München; Gesellschaft mit beschränkter Haftung; Geschäftsführer: Thomas Heurung, Frank Thürauf; Sitz der Gesellschaft: München; Registergericht München, HRB 106955
From ae14ccbd050d0b49073d5ea09de3e2af63f8c674 Mon Sep 17 00:00:00 2001 From: Thomas Schwinge <thomas@codesourcery.com> Date: Thu, 7 Jul 2022 09:45:42 +0200 Subject: [PATCH] Enhance 'libgomp.c-c++-common/requires-4.c', 'libgomp.c-c++-common/requires-5.c' testing These should compile and link and execute in all configurations; host-fallback execution, which we may actually verify. Follow-up to recent commit 683f11843974f0bdf42f79cdcbb0c2b43c7b81b0 "OpenMP: Move omp requires checks to libgomp". libgomp/ * testsuite/libgomp.c-c++-common/requires-4.c: Enhance testing. * testsuite/libgomp.c-c++-common/requires-5.c: Likewise. --- .../libgomp.c-c++-common/requires-4.c | 17 ++++++++++++----- .../libgomp.c-c++-common/requires-5.c | 18 +++++++++++------- 2 files changed, 23 insertions(+), 12 deletions(-) diff --git a/libgomp/testsuite/libgomp.c-c++-common/requires-4.c b/libgomp/testsuite/libgomp.c-c++-common/requires-4.c index 128fdbb8463..deb04368108 100644 --- a/libgomp/testsuite/libgomp.c-c++-common/requires-4.c +++ b/libgomp/testsuite/libgomp.c-c++-common/requires-4.c @@ -1,22 +1,29 @@ -/* { dg-do link { target offloading_enabled } } */ /* { dg-additional-options "-flto" } */ /* { dg-additional-sources requires-4-aux.c } */ -/* Check diagnostic by device-compiler's or host compiler's lto1. +/* Check no diagnostic by device-compiler's or host compiler's lto1. Other file uses: 'requires reverse_offload', but that's inactive as there are no declare target directives, device constructs nor device routines */ +/* For actual offload execution, prints the following (only) if GOMP_DEBUG=1: + "devices present but 'omp requires unified_address, unified_shared_memory, reverse_offload' cannot be fulfilled" + and does host-fallback execution. */ + #pragma omp requires unified_address,unified_shared_memory -int a[10]; +int a[10] = { 0 }; extern void foo (void); int main (void) { - #pragma omp target + #pragma omp target map(to: a) + for (int i = 0; i < 10; i++) + a[i] = i; + for (int i = 0; i < 10; i++) - a[i] = 0; + if (a[i] != i) + __builtin_abort (); foo (); return 0; diff --git a/libgomp/testsuite/libgomp.c-c++-common/requires-5.c b/libgomp/testsuite/libgomp.c-c++-common/requires-5.c index c1e5540cfc5..68816314b94 100644 --- a/libgomp/testsuite/libgomp.c-c++-common/requires-5.c +++ b/libgomp/testsuite/libgomp.c-c++-common/requires-5.c @@ -1,21 +1,25 @@ -/* { dg-do run { target { offload_target_nvptx || offload_target_amdgcn } } } */ /* { dg-additional-sources requires-5-aux.c } */ +/* For actual offload execution, prints the following (only) if GOMP_DEBUG=1: + "devices present but 'omp requires unified_address, unified_shared_memory, reverse_offload' cannot be fulfilled" + and does host-fallback execution. */ + #pragma omp requires unified_shared_memory, unified_address, reverse_offload -int a[10]; +int a[10] = { 0 }; extern void foo (void); int main (void) { - #pragma omp target + #pragma omp target map(to: a) + for (int i = 0; i < 10; i++) + a[i] = i; + for (int i = 0; i < 10; i++) - a[i] = 0; + if (a[i] != i) + __builtin_abort (); foo (); return 0; } - -/* (Only) if GOMP_DEBUG=1, should print at runtime the following: - "devices present but 'omp requires unified_address, unified_shared_memory, reverse_offload' cannot be fulfilled" */ -- 2.35.1