public inbox for gcc-cvs@sourceware.org
help / color / mirror / Atom feed
* [gcc r12-3058] Make the OpenMP 'error' directive work for nvptx offloading
@ 2021-08-22 9:11 Thomas Schwinge
0 siblings, 0 replies; only message in thread
From: Thomas Schwinge @ 2021-08-22 9:11 UTC (permalink / raw)
To: gcc-cvs
https://gcc.gnu.org/g:a5416bf369419428fb139c432bcd88f6f8ee4910
commit r12-3058-ga5416bf369419428fb139c432bcd88f6f8ee4910
Author: Thomas Schwinge <thomas@codesourcery.com>
Date: Fri Aug 20 15:12:56 2021 +0200
Make the OpenMP 'error' directive work for nvptx offloading
... and add a minimum amount of offloading testing.
(Leaving aside that 'fwrite' to 'stderr' probably wouldn't work anyway) the
'fwrite' calls in 'libgomp/error.c:GOMP_warning', 'libgomp/error.c:GOMP_error'
drag in 'isatty', which isn't provided by my nvptx newlib build at present, so
we get, for example:
[...]
FAIL: libgomp.c/../libgomp.c-c++-common/declare_target-1.c (test for excess errors)
Excess errors:
unresolved symbol isatty
mkoffload: fatal error: [...]/build-gcc/./gcc/x86_64-pc-linux-gnu-accel-nvptx-none-gcc returned 1 exit status
[...]
..., and many more.
Fix up for recent commit 0d973c0a0d90a0a302e7eda1a4d9709be3c5b102
"openmp: Implement the error directive".
libgomp/
* config/nvptx/error.c (fwrite, exit): Override, too.
* testsuite/libgomp.c-c++-common/error-1.c: Add a minimum amount
of offloading testing.
* testsuite/libgomp.fortran/error-1.f90: Likewise.
Diff:
---
libgomp/config/nvptx/error.c | 32 +++++++++++++++++++++---
libgomp/testsuite/libgomp.c-c++-common/error-1.c | 10 ++++++++
libgomp/testsuite/libgomp.fortran/error-1.f90 | 9 +++++++
3 files changed, 48 insertions(+), 3 deletions(-)
diff --git a/libgomp/config/nvptx/error.c b/libgomp/config/nvptx/error.c
index dfa75da354f..c55791e34b4 100644
--- a/libgomp/config/nvptx/error.c
+++ b/libgomp/config/nvptx/error.c
@@ -31,12 +31,38 @@
#include <stdio.h>
#include <stdlib.h>
-#undef vfprintf
-#undef fputs
-#undef fputc
+/* No 'FILE *stream's, just basic 'vprintf' etc. */
+
+#undef vfprintf
#define vfprintf(stream, fmt, list) vprintf (fmt, list)
+
+#undef fputs
#define fputs(s, stream) printf ("%s", s)
+
+#undef fputc
#define fputc(c, stream) printf ("%c", c)
+#undef fwrite
+#if 0
+# define fwrite(ptr, size, nmemb, stream) \
+ printf ("%.*s", (int) (size * nmemb), (int) (size * nmemb), ptr)
+/* ... prints literal '%.*s'. */
+#else
+# define fwrite(ptr, size, nmemb, stream) \
+ do { \
+ /* Yuck! */ \
+ for (size_t i = 0; i < size * nmemb; ++i) \
+ printf ("%c", ptr[i]); \
+ } while (0)
+#endif
+
+
+/* The 'exit (EXIT_FAILURE);' of an Fortran (only, huh?) OpenMP 'error'
+ directive with 'severity (fatal)' causes a hang, so 'abort' instead of
+ 'exit'. */
+#undef exit
+#define exit(status) abort ()
+
+
#include "../../error.c"
diff --git a/libgomp/testsuite/libgomp.c-c++-common/error-1.c b/libgomp/testsuite/libgomp.c-c++-common/error-1.c
index 5f454c1adaa..04c0519bf63 100644
--- a/libgomp/testsuite/libgomp.c-c++-common/error-1.c
+++ b/libgomp/testsuite/libgomp.c-c++-common/error-1.c
@@ -34,11 +34,20 @@ foo (int i, int x)
int
main ()
{
+ /* Initialize offloading early, so that any output this may produce doesn't
+ disturb the 'dg-output' scanning below. */
+ #pragma omp target
+ ;
+
if (foo (5, 0) != 13 || foo (6, 1) != 17)
abort ();
#pragma omp error at (execution) severity (warning)
const char *msg = "my message" + 2;
#pragma omp error at (execution) severity (warning) message (msg + 1)
+ #pragma omp target
+ {
+ #pragma omp error at (execution) severity (warning) message ("hello from a distance")
+ }
#pragma omp error at (execution) severity (fatal) message (msg - 2)
#pragma omp error at (execution) severity (warning) message ("foobar")
return 0;
@@ -46,4 +55,5 @@ main ()
/* { dg-output "libgomp: error directive encountered(\n|\r|\n\r)(\n|\r|\n\r)" } */
/* { dg-output "libgomp: error directive encountered: message(\n|\r|\n\r)(\n|\r|\n\r)" } */
+/* { dg-output "libgomp: error directive encountered: hello from a distance(\n|\r|\n\r)(\n|\r|\n\r)" } */
/* { dg-output "libgomp: fatal error: error directive encountered: my message" } */
diff --git a/libgomp/testsuite/libgomp.fortran/error-1.f90 b/libgomp/testsuite/libgomp.fortran/error-1.f90
index 92c246cfcaf..7c497fd002e 100644
--- a/libgomp/testsuite/libgomp.fortran/error-1.f90
+++ b/libgomp/testsuite/libgomp.fortran/error-1.f90
@@ -37,6 +37,11 @@ program main
character(len=13) :: msg
character(len=:), allocatable :: msg2, msg3
+ ! Initialize offloading early, so that any output this may produce doesn't
+ ! disturb the 'dg-output' scanning below.
+ !$omp target
+ !$omp end target
+
msg = "my message"
if (foo (5, 0) /= 15 .or. foo (7, 1) /= 16) &
stop 1
@@ -47,6 +52,9 @@ program main
!$omp error at (execution) severity (warning)
!$omp error at (execution) severity (warning) message(trim(msg(4:)))
!$omp error at (execution) severity (warning) message ("Farewell")
+ !$omp target
+ !$omp error at (execution) severity (warning) message ("ffrom a distanceee"(2:16))
+ !$omp end target
!$omp error at (execution) severity (warning) message (msg2)
!$omp error at (execution) severity (warning) message (msg(4:6))
!$omp error at (execution) severity (fatal) message (msg)
@@ -73,6 +81,7 @@ end
! { dg-output "libgomp: error directive encountered(\n|\r|\n\r)(\n|\r|\n\r)" }
! { dg-output "libgomp: error directive encountered: message(\n|\r|\n\r)(\n|\r|\n\r)" }
! { dg-output "libgomp: error directive encountered: Farewell(\n|\r|\n\r)(\n|\r|\n\r)" }
+! { dg-output "libgomp: error directive encountered: from a distance(\n|\r|\n\r)(\n|\r|\n\r)" }
! { dg-output "libgomp: error directive encountered: Hello World(\n|\r|\n\r)(\n|\r|\n\r)" }
! { dg-output "libgomp: error directive encountered: mes(\n|\r|\n\r)(\n|\r|\n\r)" }
! { dg-output "libgomp: fatal error: error directive encountered: my message (\n|\r|\n\r)" }
^ permalink raw reply [flat|nested] only message in thread
only message in thread, other threads:[~2021-08-22 9:11 UTC | newest]
Thread overview: (only message) (download: mbox.gz / follow: Atom feed)
-- links below jump to the message on this page --
2021-08-22 9:11 [gcc r12-3058] Make the OpenMP 'error' directive work for nvptx offloading 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).