[Patch] libgomp: Fix on_device_arch.c aux-file handling [PR99555] (was: [WIP] Re: [PATCH] openmp: Fix intermittent hanging of task-detach-6 libgomp tests [PR98738])

[Tobias Burnus <tobias@codesourcery.com>]

[-- Attachment #1: Type: text/plain, Size: 2547 bytes --]

Hi Thomas, hi all,

your commit causes compile fails:

cc1: fatal error: ../lib/on_device_arch.c: No such file or directory

FAIL: libgomp.c/../libgomp.c-c++-common/task-detach-6.c (test for excess errors)
FAIL: libgomp.c/pr99555-1.c (test for excess errors)
FAIL: libgomp.fortran/task-detach-6.f90   -O0  (test for excess errors)

That's with embedded testing, where files are copied into the test directory, i.e.
cp .../libgomp/testsuite/libgomp.fortran/../lib/on_device_arch.c on_device_arch.c
cp .../libgomp/testsuite/libgomp.fortran/task-detach-6.f90 task-detach-6.f90
and then executed as:
powerpc64le-none-linux-gnu-gcc $TESTDIR/task-detach-6.f90 ../lib/on_device_arch.c
which fails.

How about the following patch? It moves the aux function to libgomp.c-c++-common/on_device_arch.c
and #includes it in the new wrapper files libgomp.{c,fortran}/on_device_arch.c.
(Based on the observation that #include with relative paths always works,
while dg-additional-sources may not, depending how the testsuite it run.)

OK? Or does anyone have a better suggestion?

Tobias

PS: The testcases still FAIL with nvptx offloading – but now at execution time.
I think that's expected, is it? (→PR99555?)
FAIL: libgomp.c/../libgomp.c-c++-common/pr96390.c execution test
FAIL: libgomp.c/../libgomp.c-c++-common/task-detach-6.c execution test
FAIL: libgomp.fortran/task-detach-6.f90   -O0  execution test
FAIL: libgomp.fortran/task-detach-6.f90   -O1  execution test
FAIL: libgomp.fortran/task-detach-6.f90   -O2  execution test
FAIL: libgomp.fortran/task-detach-6.f90   -O3 -fomit-frame-pointer -funroll-loops -fpeel-loops -ftracer -finline-functions  execution test
FAIL: libgomp.fortran/task-detach-6.f90   -O3 -g  execution test
FAIL: libgomp.fortran/task-detach-6.f90   -Os  execution test

On 25.03.21 13:02, Thomas Schwinge wrote:
>> Until this gets resolved properly, OK to push something like the attached
>> (currently testing) "Avoid OpenMP/nvptx execution-time hangs for simple
>> nested OpenMP 'target'/'parallel'/'task' constructs [PR99555]"?
> [...] I've now pushed "Avoid OpenMP/nvptx execution-time hangs for
> simple nested OpenMP 'target'/'parallel'/'task' constructs [PR99555]" to
> master branch in commit d99111fd8e12deffdd9a965ce17e8a760d531ec3, see
> attached.  "... awaiting proper resolution, of course."
-----------------
Mentor Graphics (Deutschland) GmbH, Arnulfstrasse 201, 80634 München Registergericht München HRB 106955, Geschäftsführer: Thomas Heurung, Frank Thürauf

[-- Attachment #2: on_arch.diff --]
[-- Type: text/x-patch, Size: 4889 bytes --]

libgomp: Fix on_device_arch.c aux-file handling [PR99555]

libgomp/ChangeLog:

	PR target/99555
        * testsuite/libgomp.c-c++-common/task-detach-6.c:
	* testsuite/libgomp.c/pr99555-1.c:
	* testsuite/libgomp.fortran/task-detach-6.f90:
	* testsuite/lib/on_device_arch.c: Removed.
	* testsuite/libgomp.c-c++-common/on_device_arch.c: New test.
	* testsuite/libgomp.c/on_device_arch.c: New test.
	* testsuite/libgomp.fortran/on_device_arch.c: New test.

 libgomp/testsuite/lib/on_device_arch.c             | 30 --------------------
 .../libgomp.c-c++-common/on_device_arch.c          | 33 ++++++++++++++++++++++
 .../testsuite/libgomp.c-c++-common/task-detach-6.c |  2 +-
 libgomp/testsuite/libgomp.c/on_device_arch.c       |  3 ++
 libgomp/testsuite/libgomp.c/pr99555-1.c            |  2 +-
 libgomp/testsuite/libgomp.fortran/on_device_arch.c |  3 ++
 .../testsuite/libgomp.fortran/task-detach-6.f90    |  2 +-
 7 files changed, 42 insertions(+), 33 deletions(-)

diff --git a/libgomp/testsuite/lib/on_device_arch.c b/libgomp/testsuite/lib/on_device_arch.c
deleted file mode 100644
index 1c0753c..0000000
--- a/libgomp/testsuite/lib/on_device_arch.c
+++ /dev/null
@@ -1,30 +0,0 @@
-#include <gomp-constants.h>
-
-/* static */ int
-device_arch_nvptx (void)
-{
-  return GOMP_DEVICE_NVIDIA_PTX;
-}
-
-#pragma omp declare variant (device_arch_nvptx) match(construct={target},device={arch(nvptx)})
-/* static */ int
-device_arch (void)
-{
-  return GOMP_DEVICE_DEFAULT;
-}
-
-static int
-on_device_arch (int d)
-{
-  int d_cur;
-  #pragma omp target map(from:d_cur)
-  d_cur = device_arch ();
-
-  return d_cur == d;
-}
-
-int
-on_device_arch_nvptx ()
-{
-  return on_device_arch (GOMP_DEVICE_NVIDIA_PTX);
-}
diff --git a/libgomp/testsuite/libgomp.c-c++-common/on_device_arch.c b/libgomp/testsuite/libgomp.c-c++-common/on_device_arch.c
new file mode 100644
index 0000000..00524b5
--- /dev/null
+++ b/libgomp/testsuite/libgomp.c-c++-common/on_device_arch.c
@@ -0,0 +1,33 @@
+/* Auxiliar file.  */
+/* { dg-do compile  { target skip-all-targets } } */
+/* Note: this file is also #included in ../libgomp.fortran/on_device_arch.c  */
+#include <gomp-constants.h>
+
+/* static */ int
+device_arch_nvptx (void)
+{
+  return GOMP_DEVICE_NVIDIA_PTX;
+}
+
+#pragma omp declare variant (device_arch_nvptx) match(construct={target},device={arch(nvptx)})
+/* static */ int
+device_arch (void)
+{
+  return GOMP_DEVICE_DEFAULT;
+}
+
+static int
+on_device_arch (int d)
+{
+  int d_cur;
+  #pragma omp target map(from:d_cur)
+  d_cur = device_arch ();
+
+  return d_cur == d;
+}
+
+int
+on_device_arch_nvptx ()
+{
+  return on_device_arch (GOMP_DEVICE_NVIDIA_PTX);
+}
diff --git a/libgomp/testsuite/libgomp.c-c++-common/task-detach-6.c b/libgomp/testsuite/libgomp.c-c++-common/task-detach-6.c
index 4a3e4a2..c88cec2 100644
--- a/libgomp/testsuite/libgomp.c-c++-common/task-detach-6.c
+++ b/libgomp/testsuite/libgomp.c-c++-common/task-detach-6.c
@@ -1,6 +1,6 @@
 /* { dg-do run } */
 
-/* { dg-additional-sources "../lib/on_device_arch.c" } */
+/* { dg-additional-sources "on_device_arch.c" } */
 extern int on_device_arch_nvptx ();
 
 #include <omp.h>
diff --git a/libgomp/testsuite/libgomp.c/on_device_arch.c b/libgomp/testsuite/libgomp.c/on_device_arch.c
new file mode 100644
index 0000000..af71103
--- /dev/null
+++ b/libgomp/testsuite/libgomp.c/on_device_arch.c
@@ -0,0 +1,3 @@
+/* Auxiliar file.  */
+/* { dg-do compile  { target skip-all-targets } } */
+#include "../libgomp.c-c++-common/on_device_arch.c"
diff --git a/libgomp/testsuite/libgomp.c/pr99555-1.c b/libgomp/testsuite/libgomp.c/pr99555-1.c
index 9ba3309..d661a888 100644
--- a/libgomp/testsuite/libgomp.c/pr99555-1.c
+++ b/libgomp/testsuite/libgomp.c/pr99555-1.c
@@ -2,7 +2,7 @@
 
 // { dg-additional-options "-O0" }
 
-// { dg-additional-sources "../lib/on_device_arch.c" }
+// { dg-additional-sources "on_device_arch.c" }
 extern int on_device_arch_nvptx ();
 
 int main (void)
diff --git a/libgomp/testsuite/libgomp.fortran/on_device_arch.c b/libgomp/testsuite/libgomp.fortran/on_device_arch.c
new file mode 100644
index 0000000..af71103
--- /dev/null
+++ b/libgomp/testsuite/libgomp.fortran/on_device_arch.c
@@ -0,0 +1,3 @@
+/* Auxiliar file.  */
+/* { dg-do compile  { target skip-all-targets } } */
+#include "../libgomp.c-c++-common/on_device_arch.c"
diff --git a/libgomp/testsuite/libgomp.fortran/task-detach-6.f90 b/libgomp/testsuite/libgomp.fortran/task-detach-6.f90
index eda20e7..bd0beb6 100644
--- a/libgomp/testsuite/libgomp.fortran/task-detach-6.f90
+++ b/libgomp/testsuite/libgomp.fortran/task-detach-6.f90
@@ -1,6 +1,6 @@
 ! { dg-do run }
 
-! { dg-additional-sources ../lib/on_device_arch.c }
+! { dg-additional-sources on_device_arch.c }
   ! { dg-prune-output "command-line option '-fintrinsic-modules-path=.*' is valid for Fortran but not for C" }
 
 ! Test tasks with detach clause on an offload device.  Each device