* [gomp4, committed] Add goacc/kernels-acc-on-device.c @ 2015-10-10 10:49 Tom de Vries 2015-10-12 10:49 ` Thomas Schwinge 0 siblings, 1 reply; 5+ messages in thread From: Tom de Vries @ 2015-10-10 10:49 UTC (permalink / raw) To: gcc-patches [-- Attachment #1: Type: text/plain, Size: 110 bytes --] Hi, this patch adds a new kernels region test-case. Committed to gomp-4_0-branch as trivial. Thanks, - Tom [-- Attachment #2: 0001-Add-goacc-kernels-acc-on-device.c.patch --] [-- Type: text/x-patch, Size: 1203 bytes --] Add goacc/kernels-acc-on-device.c 2015-10-10 Tom de Vries <tom@codesourcery.com> * c-c++-common/goacc/kernels-acc-on-device.c: New test. --- .../c-c++-common/goacc/kernels-acc-on-device.c | 39 ++++++++++++++++++++++ 1 file changed, 39 insertions(+) create mode 100644 gcc/testsuite/c-c++-common/goacc/kernels-acc-on-device.c diff --git a/gcc/testsuite/c-c++-common/goacc/kernels-acc-on-device.c b/gcc/testsuite/c-c++-common/goacc/kernels-acc-on-device.c new file mode 100644 index 0000000..e9e93c7 --- /dev/null +++ b/gcc/testsuite/c-c++-common/goacc/kernels-acc-on-device.c @@ -0,0 +1,39 @@ +/* { dg-additional-options "-O2" } */ + +#include <openacc.h> + +#define N 32 + +void +foo (float *a, float *b) +{ + float exp; + int i; + int n; + +#pragma acc kernels copyin(a[0:N]) copyout(b[0:N]) + { + int ii; + + for (ii = 0; ii < N; ii++) + { + if (acc_on_device (acc_device_host)) + b[ii] = a[ii] + 1; + else + b[ii] = a[ii]; + } + } + +#pragma acc kernels copyin(a[0:N]) copyout(b[0:N]) + { + int ii; + + for (ii = 0; ii < N; ii++) + { + if (acc_on_device (acc_device_host)) + b[ii] = a[ii] + 2; + else + b[ii] = a[ii]; + } + } +} -- 1.9.1 ^ permalink raw reply [flat|nested] 5+ messages in thread
* Re: [gomp4, committed] Add goacc/kernels-acc-on-device.c 2015-10-10 10:49 [gomp4, committed] Add goacc/kernels-acc-on-device.c Tom de Vries @ 2015-10-12 10:49 ` Thomas Schwinge 2015-10-12 12:53 ` Tom de Vries 0 siblings, 1 reply; 5+ messages in thread From: Thomas Schwinge @ 2015-10-12 10:49 UTC (permalink / raw) To: Tom de Vries; +Cc: gcc-patches [-- Attachment #1: Type: text/plain, Size: 1525 bytes --] Hi Tom! On Sat, 10 Oct 2015 12:49:01 +0200, Tom de Vries <Tom_deVries@mentor.com> wrote: > --- /dev/null > +++ b/gcc/testsuite/c-c++-common/goacc/kernels-acc-on-device.c > @@ -0,0 +1,39 @@ > +/* { dg-additional-options "-O2" } */ > + > +#include <openacc.h> That doesn't work (at least in build-tree testing), as gcc/testsuite/ is not set up to look for header files in [target]/libgomp/: [...]/source-gcc/gcc/testsuite/c-c++-common/goacc/kernels-acc-on-device.c:3:21: fatal error: openacc.h: No such file or directory compilation terminated. compiler exited with status 1 > + > +#define N 32 > + > +void > +foo (float *a, float *b) > +{ > + float exp; > + int i; > + int n; > + > +#pragma acc kernels copyin(a[0:N]) copyout(b[0:N]) > + { > + int ii; > + > + for (ii = 0; ii < N; ii++) > + { > + if (acc_on_device (acc_device_host)) Your two options are: if that's applicable/sufficient for what you intend to test here, use __builtin_acc_on_device with a hard-coded acc_device_*, or duplicate part of <openacc.h> as done for example in gcc/testsuite/c-c++-common/goacc/acc_on_device-2.c. > + b[ii] = a[ii] + 1; > + else > + b[ii] = a[ii]; > + } > + } > + > +#pragma acc kernels copyin(a[0:N]) copyout(b[0:N]) > + { > + int ii; > + > + for (ii = 0; ii < N; ii++) > + { > + if (acc_on_device (acc_device_host)) > + b[ii] = a[ii] + 2; > + else > + b[ii] = a[ii]; > + } > + } > +} Grüße, Thomas [-- Attachment #2: signature.asc --] [-- Type: application/pgp-signature, Size: 472 bytes --] ^ permalink raw reply [flat|nested] 5+ messages in thread
* Re: [gomp4, committed] Add goacc/kernels-acc-on-device.c 2015-10-12 10:49 ` Thomas Schwinge @ 2015-10-12 12:53 ` Tom de Vries 2015-10-13 15:50 ` Tom de Vries 0 siblings, 1 reply; 5+ messages in thread From: Tom de Vries @ 2015-10-12 12:53 UTC (permalink / raw) To: Thomas Schwinge; +Cc: gcc-patches [-- Attachment #1: Type: text/plain, Size: 1513 bytes --] On 12/10/15 12:49, Thomas Schwinge wrote: > Hi Tom! > > On Sat, 10 Oct 2015 12:49:01 +0200, Tom de Vries<Tom_deVries@mentor.com> wrote: >> >--- /dev/null >> >+++ b/gcc/testsuite/c-c++-common/goacc/kernels-acc-on-device.c >> >@@ -0,0 +1,39 @@ >> >+/* { dg-additional-options "-O2" } */ >> >+ >> >+#include <openacc.h> Hi Thomas, > That doesn't work (at least in build-tree testing), as gcc/testsuite/ is > not set up to look for header files in [target]/libgomp/: > > [...]/source-gcc/gcc/testsuite/c-c++-common/goacc/kernels-acc-on-device.c:3:21: fatal error: openacc.h: No such file or directory > compilation terminated. > compiler exited with status 1 > Ah, I see. I was doing 'make' followed by 'make install', and then build-tree testing. The build-tree testing seems to pick up the header file from the install directory. So for me test passed. >> >+ >> >+#define N 32 >> >+ >> >+void >> >+foo (float *a, float *b) >> >+{ >> >+ float exp; >> >+ int i; >> >+ int n; >> >+ >> >+#pragma acc kernels copyin(a[0:N]) copyout(b[0:N]) >> >+ { >> >+ int ii; >> >+ >> >+ for (ii = 0; ii < N; ii++) >> >+ { >> >+ if (acc_on_device (acc_device_host)) > Your two options are: if that's applicable/sufficient for what you intend > to test here, use __builtin_acc_on_device with a hard-coded acc_device_*, > or duplicate part of <openacc.h> as done for example in > gcc/testsuite/c-c++-common/goacc/acc_on_device-2.c. > Went with second option, committed as attached. Thanks, - Tom [-- Attachment #2: 0001-Remove-openacc.h-include-from-goacc-kernels-acc-on-d.patch --] [-- Type: text/x-patch, Size: 1705 bytes --] Remove openacc.h include from goacc/kernels-acc-on-device.c 2015-10-12 Tom de Vries <tom@codesourcery.com> * c-c++-common/goacc/kernels-acc-on-device.c: Remove openacc.h include. (enum acc_device_t, acc_on_device): Declare. (foo): Remove unused vars. Use acc_device_X. --- .../c-c++-common/goacc/kernels-acc-on-device.c | 27 ++++++++++++++++------ 1 file changed, 20 insertions(+), 7 deletions(-) diff --git a/gcc/testsuite/c-c++-common/goacc/kernels-acc-on-device.c b/gcc/testsuite/c-c++-common/goacc/kernels-acc-on-device.c index e9e93c7..784c66a 100644 --- a/gcc/testsuite/c-c++-common/goacc/kernels-acc-on-device.c +++ b/gcc/testsuite/c-c++-common/goacc/kernels-acc-on-device.c @@ -1,23 +1,36 @@ /* { dg-additional-options "-O2" } */ -#include <openacc.h> +#if __cplusplus +extern "C" { +#endif + +#if __cplusplus >= 201103 +# define __GOACC_NOTHROW noexcept +#elif __cplusplus +# define __GOACC_NOTHROW throw () +#else /* Not C++ */ +# define __GOACC_NOTHROW __attribute__ ((__nothrow__)) +#endif + +typedef enum acc_device_t { acc_device_X = 123 } acc_device_t; +int acc_on_device (int) __GOACC_NOTHROW; + +#if __cplusplus +} +#endif #define N 32 void foo (float *a, float *b) { - float exp; - int i; - int n; - #pragma acc kernels copyin(a[0:N]) copyout(b[0:N]) { int ii; for (ii = 0; ii < N; ii++) { - if (acc_on_device (acc_device_host)) + if (acc_on_device (acc_device_X)) b[ii] = a[ii] + 1; else b[ii] = a[ii]; @@ -30,7 +43,7 @@ foo (float *a, float *b) for (ii = 0; ii < N; ii++) { - if (acc_on_device (acc_device_host)) + if (acc_on_device (acc_device_X)) b[ii] = a[ii] + 2; else b[ii] = a[ii]; -- 1.9.1 ^ permalink raw reply [flat|nested] 5+ messages in thread
* Re: [gomp4, committed] Add goacc/kernels-acc-on-device.c 2015-10-12 12:53 ` Tom de Vries @ 2015-10-13 15:50 ` Tom de Vries 2015-10-30 8:32 ` Thomas Schwinge 0 siblings, 1 reply; 5+ messages in thread From: Tom de Vries @ 2015-10-13 15:50 UTC (permalink / raw) To: Thomas Schwinge; +Cc: gcc-patches [-- Attachment #1: Type: text/plain, Size: 1745 bytes --] On 12/10/15 14:52, Tom de Vries wrote: > On 12/10/15 12:49, Thomas Schwinge wrote: >> Hi Tom! >> >> On Sat, 10 Oct 2015 12:49:01 +0200, Tom de >> Vries<Tom_deVries@mentor.com> wrote: >>> >--- /dev/null >>> >+++ b/gcc/testsuite/c-c++-common/goacc/kernels-acc-on-device.c >>> >@@ -0,0 +1,39 @@ >>> >+/* { dg-additional-options "-O2" } */ >>> >+ >>> >+#include <openacc.h> > > Hi Thomas, > >> That doesn't work (at least in build-tree testing), as gcc/testsuite/ is >> not set up to look for header files in [target]/libgomp/: >> >> [...]/source-gcc/gcc/testsuite/c-c++-common/goacc/kernels-acc-on-device.c:3:21: >> fatal error: openacc.h: No such file or directory >> compilation terminated. >> compiler exited with status 1 >> > > Ah, I see. I was doing 'make' followed by 'make install', and then > build-tree testing. The build-tree testing seems to pick up the header > file from the install directory. So for me test passed. > >>> >+ >>> >+#define N 32 >>> >+ >>> >+void >>> >+foo (float *a, float *b) >>> >+{ >>> >+ float exp; >>> >+ int i; >>> >+ int n; >>> >+ >>> >+#pragma acc kernels copyin(a[0:N]) copyout(b[0:N]) >>> >+ { >>> >+ int ii; >>> >+ >>> >+ for (ii = 0; ii < N; ii++) >>> >+ { >>> >+ if (acc_on_device (acc_device_host)) >> Your two options are: if that's applicable/sufficient for what you intend >> to test here, use __builtin_acc_on_device with a hard-coded acc_device_*, >> or duplicate part of <openacc.h> as done for example in >> gcc/testsuite/c-c++-common/goacc/acc_on_device-2.c. >> > > Went with second option, committed as attached. As a follow-up patch, I've factored the code into a mockup openacc.h, now shared by several test-cases. Committed to gomp-4_0-branch. Thanks, - Tom [-- Attachment #2: 0001-Factor-out-goacc-openacc.h.patch --] [-- Type: text/x-patch, Size: 3405 bytes --] Factor out goacc/openacc.h 2015-10-13 Tom de Vries <tom@codesourcery.com> * c-c++-common/goacc/openacc.h: New header file, factored out of ... * c-c++-common/goacc/kernels-acc-on-device.c: ... here. * c-c++-common/goacc/acc_on_device-2-off.c: Use openacc.h. * c-c++-common/goacc/acc_on_device-2.c: Same. --- .../c-c++-common/goacc/acc_on_device-2-off.c | 11 +---------- gcc/testsuite/c-c++-common/goacc/acc_on_device-2.c | 13 +------------ .../c-c++-common/goacc/kernels-acc-on-device.c | 19 +------------------ gcc/testsuite/c-c++-common/goacc/openacc.h | 18 ++++++++++++++++++ 4 files changed, 21 insertions(+), 40 deletions(-) create mode 100644 gcc/testsuite/c-c++-common/goacc/openacc.h diff --git a/gcc/testsuite/c-c++-common/goacc/acc_on_device-2-off.c b/gcc/testsuite/c-c++-common/goacc/acc_on_device-2-off.c index 71abe11..cce58de 100644 --- a/gcc/testsuite/c-c++-common/goacc/acc_on_device-2-off.c +++ b/gcc/testsuite/c-c++-common/goacc/acc_on_device-2-off.c @@ -3,16 +3,7 @@ /* Duplicate parts of libgomp/openacc.h, because we can't include it here. */ -#if __cplusplus -extern "C" { -#endif - -typedef enum acc_device_t { acc_device_X = 123 } acc_device_t; -extern int acc_on_device (int); - -#if __cplusplus -} -#endif +#include "openacc.h" int f (void) diff --git a/gcc/testsuite/c-c++-common/goacc/acc_on_device-2.c b/gcc/testsuite/c-c++-common/goacc/acc_on_device-2.c index 243e562..19a5bd3 100644 --- a/gcc/testsuite/c-c++-common/goacc/acc_on_device-2.c +++ b/gcc/testsuite/c-c++-common/goacc/acc_on_device-2.c @@ -1,18 +1,7 @@ /* Have to enable optimizations, as otherwise builtins won't be expanded. */ /* { dg-additional-options "-O -fdump-rtl-expand" } */ -/* Duplicate parts of libgomp/openacc.h, because we can't include it here. */ - -#if __cplusplus -extern "C" { -#endif - -typedef enum acc_device_t { acc_device_X = 123 } acc_device_t; -extern int acc_on_device (int); - -#if __cplusplus -} -#endif +#include "openacc.h" int f (void) diff --git a/gcc/testsuite/c-c++-common/goacc/kernels-acc-on-device.c b/gcc/testsuite/c-c++-common/goacc/kernels-acc-on-device.c index 784c66a..958b65b 100644 --- a/gcc/testsuite/c-c++-common/goacc/kernels-acc-on-device.c +++ b/gcc/testsuite/c-c++-common/goacc/kernels-acc-on-device.c @@ -1,23 +1,6 @@ /* { dg-additional-options "-O2" } */ -#if __cplusplus -extern "C" { -#endif - -#if __cplusplus >= 201103 -# define __GOACC_NOTHROW noexcept -#elif __cplusplus -# define __GOACC_NOTHROW throw () -#else /* Not C++ */ -# define __GOACC_NOTHROW __attribute__ ((__nothrow__)) -#endif - -typedef enum acc_device_t { acc_device_X = 123 } acc_device_t; -int acc_on_device (int) __GOACC_NOTHROW; - -#if __cplusplus -} -#endif +#include "openacc.h" #define N 32 diff --git a/gcc/testsuite/c-c++-common/goacc/openacc.h b/gcc/testsuite/c-c++-common/goacc/openacc.h new file mode 100644 index 0000000..a74a482 --- /dev/null +++ b/gcc/testsuite/c-c++-common/goacc/openacc.h @@ -0,0 +1,18 @@ +#if __cplusplus +extern "C" { +#endif + +#if __cplusplus >= 201103 +# define __GOACC_NOTHROW noexcept +#elif __cplusplus +# define __GOACC_NOTHROW throw () +#else /* Not C++ */ +# define __GOACC_NOTHROW __attribute__ ((__nothrow__)) +#endif + +typedef enum acc_device_t { acc_device_X = 123 } acc_device_t; +int acc_on_device (int) __GOACC_NOTHROW; + +#if __cplusplus +} +#endif -- 1.9.1 ^ permalink raw reply [flat|nested] 5+ messages in thread
* Re: [gomp4, committed] Add goacc/kernels-acc-on-device.c 2015-10-13 15:50 ` Tom de Vries @ 2015-10-30 8:32 ` Thomas Schwinge 0 siblings, 0 replies; 5+ messages in thread From: Thomas Schwinge @ 2015-10-30 8:32 UTC (permalink / raw) To: Tom de Vries; +Cc: gcc-patches, Nathan Sidwell [-- Attachment #1: Type: text/plain, Size: 2080 bytes --] Hi Tom! On Tue, 13 Oct 2015 17:49:21 +0200, Tom de Vries <Tom_deVries@mentor.com> wrote: > On 12/10/15 14:52, Tom de Vries wrote: > > On 12/10/15 12:49, Thomas Schwinge wrote: > >> On Sat, 10 Oct 2015 12:49:01 +0200, Tom de > >> Vries<Tom_deVries@mentor.com> wrote: > >>> >--- /dev/null > >>> >+++ b/gcc/testsuite/c-c++-common/goacc/kernels-acc-on-device.c > >>> >@@ -0,0 +1,39 @@ > >>> >+/* { dg-additional-options "-O2" } */ > >>> >+ > >>> >+#include <openacc.h> > > > > Hi Thomas, > > > >> That doesn't work (at least in build-tree testing), as gcc/testsuite/ is > >> not set up to look for header files in [target]/libgomp/: > >> > >> [...]/source-gcc/gcc/testsuite/c-c++-common/goacc/kernels-acc-on-device.c:3:21: > >> fatal error: openacc.h: No such file or directory > >> compilation terminated. > >> compiler exited with status 1 > As a follow-up patch, I've factored the code into a mockup openacc.h, > now shared by several test-cases. > Factor out goacc/openacc.h > > 2015-10-13 Tom de Vries <tom@codesourcery.com> > > * c-c++-common/goacc/openacc.h: New header file, factored out of ... > * c-c++-common/goacc/kernels-acc-on-device.c: ... here. > * c-c++-common/goacc/acc_on_device-2-off.c: Use openacc.h. > * c-c++-common/goacc/acc_on_device-2.c: Same. As a clean-up, and to move acc_on_device testing to where that OpenACC Runtime Library routine is actually defined (libgomp's openacc.h), Nathan has just removed this stub openacc.h header file, and also the test files listed just above, plus the gcc/testsuite/c-c++-common/goacc/kernels-acc-on-device-2.c that you added later, see: <http://news.gmane.org/find-root.php?message_id=%3C56327E51.1090507%40acm.org%3E> (gomp-4_0-branch), <http://news.gmane.org/find-root.php?message_id=%3C5632B856.4050509%40acm.org%3E> (trunk). If the kernels tests are still important (they tested ICEs, as far as I remember), you'll have to re-instantiate these in libgomp/testsuite/, or using __builtin_acc_on_device in gcc/testsuite/. Grüße Thomas [-- Attachment #2: signature.asc --] [-- Type: application/pgp-signature, Size: 472 bytes --] ^ permalink raw reply [flat|nested] 5+ messages in thread
end of thread, other threads:[~2015-10-30 8:01 UTC | newest] Thread overview: 5+ messages (download: mbox.gz / follow: Atom feed) -- links below jump to the message on this page -- 2015-10-10 10:49 [gomp4, committed] Add goacc/kernels-acc-on-device.c Tom de Vries 2015-10-12 10:49 ` Thomas Schwinge 2015-10-12 12:53 ` Tom de Vries 2015-10-13 15:50 ` Tom de Vries 2015-10-30 8:32 ` Thomas Schwinge
This is a public inbox, see mirroring instructions for how to clone and mirror all data and code used for this inbox; as well as URLs for read-only IMAP folder(s) and NNTP newsgroup(s).