Re: [PATCH] xfail and improve some failing libgomp tests

[Tom de Vries]

On 10/5/20 3:15 PM, Tom de Vries wrote:
> On 2/7/20 4:29 PM, Jakub Jelinek wrote:
>> On Fri, Feb 07, 2020 at 09:56:38AM +0100, Harwath, Frederik wrote:
>>> * {target-32.c, thread-limit-2.c}:
>>> no "usleep" implemented for nvptx. Cf. https://gcc.gnu.org/PR81690
>>
>> Please don't, I want to deal with that using declare variant, just didn't
>> get yet around to finishing the last patch needed for that.  Will try next 
>> week.
>>
> 
> Hi Jakub,
> 
> Ping, any update on this?

FWIW, I've tried as in patch attached below, but I didn't get it
compiling, I still got:
...
FAIL: libgomp.c/target-32.c (test for excess errors)
Excess errors:
unresolved symbol usleep
...

Jakub, is this already supposed to work?

Thanks,
- Tom

diff --git a/libgomp/testsuite/libgomp.c/target-32.c b/libgomp/testsuite/libgomp.c/target-32.c
index 233877b702b..7ddf8721ed3 100644
--- a/libgomp/testsuite/libgomp.c/target-32.c
+++ b/libgomp/testsuite/libgomp.c/target-32.c
@@ -1,6 +1,26 @@
 #include <stdlib.h>
 #include <unistd.h>
 
+extern void base_delay(int);
+extern void nvptx_delay(int);
+
+#pragma omp declare variant( nvptx_delay ) match( construct={target}, implementation={vendor(nvidia)} )
+void base_delay(int d)
+{
+  usleep (d);
+}
+
+void nvptx_delay(int 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");
+}
+
 int main ()
 {
   int a = 0, b = 0, c = 0, d[7];
@@ -18,28 +38,28 @@ int main ()
 
     #pragma omp target nowait map(alloc: b) depend(in: d[2]) depend(out: d[3])
     {
-      usleep (1000);
+      base_delay (1000);
       #pragma omp atomic update
       b |= 4;
     }
 
     #pragma omp target nowait map(alloc: b) depend(in: d[2]) depend(out: d[4])
     {
-      usleep (5000);
+      base_delay (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);
+      base_delay (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);
+      base_delay (1000);
       #pragma omp atomic update
       c |= 2;
     }