On Thu, Jul 30, 2015 at 10:12:59 +0200, Jakub Jelinek wrote:
> This test will fail on HSA, you don't assume just that it doesn't
> fallback to host, but also non-shared address space.
> I think it would be better to start with some check for non-shared address
> space, like:
> /* This test relies on non-shared address space. Punt otherwise. */
> void ensure_nonshared_as (void)
> {
> int a = 8;
> #pragma omp target map(to:a)
> {
> a++;
> }
> if (a == 8)
> exit (0);
> }
>
> And generally, it is better to have most of the tests not relying on
> offloading only or even non-shared address space, so that we also test
> shared address space and host fallback. But a few tests won't hurt...
Sure, but it's not possible to fully test data mapping without non-shared
address space. I've created new check_effective_target, ok for gomp-4_1-branch?
* testsuite/lib/libgomp.exp
(check_effective_target_offload_device_nonshared_as): New.
* testsuite/libgomp.c++/examples-4/e.53.2.C: Replace offload_device with
offload_device_nonshared_as.
* testsuite/libgomp.c/target-11.c: Ditto.
diff --git a/libgomp/testsuite/lib/libgomp.exp
b/libgomp/testsuite/lib/libgomp.exp
index 438777f..3a29b78 100644
--- a/libgomp/testsuite/lib/libgomp.exp
+++ b/libgomp/testsuite/lib/libgomp.exp
@@ -320,6 +320,19 @@ proc check_effective_target_offload_device { } {
} ]
}
+# Return 1 if offload device is available and it has non-shared address space.
+proc check_effective_target_offload_device_nonshared_as { } {
+ return [check_runtime_nocache offload_device_nonshared_as {
+ int main ()
+ {
+ int a = 8;
+ #pragma omp target map(to: a)
+ a++;
+ return a != 8;
+ }
+ } ]
+}
+
# Return 1 if at least one nvidia board is present.
proc check_effective_target_openacc_nvidia_accel_present { } {
diff --git a/libgomp/testsuite/libgomp.c++/examples-4/e.53.2.C
b/libgomp/testsuite/libgomp.c++/examples-4/e.53.2.C
index 75276e7..6d5b5e4 100644
--- a/libgomp/testsuite/libgomp.c++/examples-4/e.53.2.C
+++ b/libgomp/testsuite/libgomp.c++/examples-4/e.53.2.C
@@ -1,5 +1,5 @@
// { dg-do run }
-// { dg-require-effective-target offload_device }
+// { dg-require-effective-target offload_device_nonshared_as }
#include <stdlib.h>
diff --git a/libgomp/testsuite/libgomp.c/target-11.c
b/libgomp/testsuite/libgomp.c/target-11.c
index b86097a..ed6a17a 100644
--- a/libgomp/testsuite/libgomp.c/target-11.c
+++ b/libgomp/testsuite/libgomp.c/target-11.c
@@ -1,4 +1,4 @@
-/* { dg-require-effective-target offload_device } */
+/* { dg-require-effective-target offload_device_nonshared_as } */
#include <stdlib.h>
#include <assert.h>
-- Ilya
