diff mbox

[gomp4.1] Support #pragma omp target {enter,exit} data

Message ID 20150730144015.GA22099@msticlxl57.ims.intel.com
State New
Headers show

Commit Message

Ilya Verbin July 30, 2015, 2:40 p.m. UTC 
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.




  -- Ilya

Comments

Jakub Jelinek July 30, 2015, 2:46 p.m. UTC | #1 
On Thu, Jul 30, 2015 at 05:40:15PM +0300, Ilya Verbin wrote:
> 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.

Ok.

	Jakub
diff mbox

Patch

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>