xfail and improve some failing libgomp tests [PR81690]

With the patch I've posted today to fix up declare variant LTO handling,
Tobias reported the patch still doesn't work, and there are two
reasons for that.
One is that when the base function is marked implicitly as declare target,
we don't mark also implicitly the variants.  I'll need to ask on omp-lang
about details for that, but generally the compiler should do it some way.
The other one is that the way base_delay is written, it will always
call the usleep function, which is undesirable for nvptx.  While the
compiler will replace all direct calls to base_delay to nvptx_delay,
the base_delay definition which calls usleep stays.

2020-10-28  Jakub Jelinek  <jakub@redhat.com>
	    Tom de Vries  <tdevries@suse.de>

	PR testsuite/81690
	* testsuite/libgomp.c/usleep.h: New file.
	* testsuite/libgomp.c/target-32.c: Include usleep.h.
	(main): Use tgt_usleep instead of usleep.
	* testsuite/libgomp.c/thread-limit-2.c: Include usleep.h.
	(main): Use tgt_usleep instead of usleep.

diff --git a/libgomp/testsuite/libgomp.c/target-32.c b/libgomp/testsuite/libgomp.c/target-32.c
index 233877b702ba6903ba640c3b2ca9e9fb425d3b2b..1444523615eaf127a913e5b8b5f628d2ce7f41f6 100644 (file)
--- a/libgomp/testsuite/libgomp.c/target-32.c
+++ b/libgomp/testsuite/libgomp.c/target-32.c
@@ -1,5 +1,6 @@
 #include <stdlib.h>
 #include <unistd.h>
+#include "usleep.h"
 
 int main ()
 {
@@ -18,28 +19,28 @@ int main ()
 
     #pragma omp target nowait map(alloc: b) depend(in: d[2]) depend(out: d[3])
     {
-      usleep (1000);
+      tgt_usleep (1000);
       #pragma omp atomic update
       b |= 4;
     }
 
     #pragma omp target nowait map(alloc: b) depend(in: d[2]) depend(out: d[4])
     {
-      usleep (5000);
+      tgt_usleep (5000);
       #pragma omp atomic update
       b |= 1;
     }
 
     #pragma omp target nowait map(alloc: c) depend(in: d[3], d[4]) depend(out: d[5])
     {
-      usleep (5000);
+      tgt_usleep (5000);
       #pragma omp atomic update
       c |= 8;
     }
 
     #pragma omp target nowait map(alloc: c) depend(in: d[3], d[4]) depend(out: d[6])
     {
-      usleep (1000);
+      tgt_usleep (1000);
       #pragma omp atomic update
       c |= 2;
     }

diff --git a/libgomp/testsuite/libgomp.c/thread-limit-2.c b/libgomp/testsuite/libgomp.c/thread-limit-2.c
index 1a97fb6298596122b041eb8270bd802e9bd85bf8..dc247a714fcf69bebe88924d24cae7299109b35b 100644 (file)
--- a/libgomp/testsuite/libgomp.c/thread-limit-2.c
+++ b/libgomp/testsuite/libgomp.c/thread-limit-2.c
@@ -4,6 +4,7 @@
 #include <stdlib.h>
 #include <unistd.h>
 #include <omp.h>
+#include "usleep.h"
 
 int
 main ()
@@ -48,7 +49,7 @@ main ()
          v = ++cnt;
          if (v > 6)
            abort ();
-         usleep (10000);
+         tgt_usleep (10000);
          #pragma omp atomic
          --cnt;
        }

diff --git a/libgomp/testsuite/libgomp.c/usleep.h b/libgomp/testsuite/libgomp.c/usleep.h
new file mode 100644 (file)
index 0000000..c01aaa0
--- /dev/null
+++ b/libgomp/testsuite/libgomp.c/usleep.h
@@ -0,0 +1,24 @@
+#include <unistd.h>
+
+int
+nvptx_usleep (useconds_t d)
+{
+  /* This function serves as a replacement for usleep in
+     this test case.  It does not even attempt to be functionally
+     equivalent  - we just want some sort of delay. */
+  int i;
+  int N = d * 2000;
+  for (i = 0; i < N; i++)
+    asm volatile ("" : : : "memory");
+  return 0;
+}
+
+#pragma omp declare variant (nvptx_usleep) match(construct={target},device={arch(nvptx)})
+#pragma omp declare variant (usleep) match(user={condition(1)})
+int
+tgt_usleep (useconds_t d)
+{
+  return 0;
+}
+
+#pragma omp declare target to (nvptx_usleep, tgt_usleep)