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.

3 files changed, 48 insertions, 3 deletions

diff --git a/libgomp/config/nvptx/error.c b/libgomp/config/nvptx/error.c
index dfa75da..c55791e 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 5f454c1..04c0519 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 92c246c..7c497fd 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)" }