New tests usm (first 5 from PR 632)

tests/5.0/requires/test_requires_unified_shared_memory_allocatable.F90

@@ -0,0 +1,76 @@
+!===--- test_requires_unified_shared_memory_allocatable.F90 ------------------------------===//
+! 
+! OpenMP API Version 5.0 Nov 2018
+!
+! This test checks for unified shared memory support of an allocatable array.
+!
+!===------------------------------------------------------------------------------===//
+
+#define OMPVV_MODULE_REQUIRES_LINE !$omp requires unified_shared_memory
+#include "ompvv.F90"
+
+#define N 1024
+
+PROGRAM test_requires_unified_shared_memory_allocatable
+   USE iso_fortran_env
+   USE ompvv_lib
+   USE omp_lib
+   implicit none
+
+!$omp requires unified_shared_memory
+
+   OMPVV_TEST_OFFLOADING
+   OMPVV_TEST_VERBOSE(unified_shared_memory_allocatable() .NE. 0)
+   OMPVV_REPORT_AND_RETURN()
+
+CONTAINS
+  INTEGER FUNCTION unified_shared_memory_allocatable()
+    INTEGER:: errors, i
+    INTEGER, ALLOCATABLE:: anArray(:)
+    INTEGER, ALLOCATABLE:: anArrayCopy(:)
+
+    OMPVV_INFOMSG("Unified shared memory testing - Array on allocatable")
+
+    errors = 0
+
+    ALLOCATE(anArray(N), anArrayCopy(N))
+
+    OMPVV_ERROR_IF(.NOT. ALLOCATED(anArray), "Memory was not properly allocated")
+    OMPVV_ERROR_IF(.NOT. ALLOCATED(anArrayCopy), "Memory was not properly allocated")
+
+    DO i = 1, N
+      anArray(i) = i
+      anArrayCopy(i) = 0
+    END DO
+
+    ! Modify in the device
+    !$omp target
+    DO i = 1, N
+      anArray(i) = anArray(i) + 10
+    END DO
+    !$omp end target
+
+    ! Modify again on the host
+    DO i = 1, N
+      anArray(i) = anArray(i) + 10
+    END DO
+
+    ! Get the value the device is seeing
+    !$omp target
+    DO i = 1, N
+      anArrayCopy(i) = anArray(i)
+    END DO
+    !$omp end target
+
+    DO i = 1, N
+      OMPVV_TEST_AND_SET_VERBOSE(errors, anArray(i) .NE. (i + 20))
+      OMPVV_TEST_AND_SET_VERBOSE(errors, anArrayCopy(i) .NE. (i + 20))
+      IF (errors .NE. 0) THEN
+        exit
+      END IF
+    END DO
+
+    DEALLOCATE(anArray, anArrayCopy)
+    unified_shared_memory_allocatable = errors
+  END FUNCTION unified_shared_memory_allocatable
+END PROGRAM test_requires_unified_shared_memory_allocatable

tests/5.0/requires/test_requires_unified_shared_memory_allocatable_map.F90

@@ -0,0 +1,73 @@
+!===--- test_requires_unified_shared_memory_allocatable_map.F90 --------------------------===//
+! 
+! OpenMP API Version 5.0 Nov 2018
+!
+! This test checks for unified shared memory of an array that is allocated  
+! using allocatable and that is accessed from host and device with the same pointer.
+!
+! The mapping of a pointer under shared memory should allow for the pointer to
+! have the same value as in the host.
+!
+!===------------------------------------------------------------------------------===//
+
+#define OMPVV_MODULE_REQUIRES_LINE !$omp requires unified_shared_memory
+#include "ompvv.F90"
+
+#define N 1024
+
+PROGRAM test_requires_unified_shared_memory_allocatable_map
+   USE iso_fortran_env
+   USE ompvv_lib
+   USE omp_lib
+   implicit none
+   LOGICAL:: isOffloading
+
+!$omp requires unified_shared_memory
+
+  OMPVV_TEST_AND_SET_OFFLOADING(isOffloading)
+
+  OMPVV_WARNING_IF(.NOT. isOffloading, "With no offloading, unified shared memory is guaranteed due to host execution")
+
+   OMPVV_TEST_VERBOSE(unified_shared_memory_allocatable_map() .NE. 0)
+
+   OMPVV_REPORT_AND_RETURN()
+
+CONTAINS
+  INTEGER FUNCTION unified_shared_memory_allocatable_map()
+    INTEGER:: errors, i
+    INTEGER, ALLOCATABLE:: anArray(:)
+    INTEGER:: ERR
+
+    OMPVV_INFOMSG("Unified shared memory testing - Array on allocatable")
+
+    errors = 0
+
+    ALLOCATE(anArray(N),  STAT=ERR)
+
+    IF( ERR /= 0 ) THEN
+      OMPVV_ERROR("Memory was not properly allocated")
+      OMPVV_RETURN(ERR)
+    END IF
+
+    DO i = 1, N
+      anArray(i) = i
+    END DO
+
+    ! Modify in the device
+    !$omp target map(anArray)
+    DO i = 1, N
+      anArray(i) = anArray(i) + 10
+    END DO
+    !$omp end target
+
+    DO i = 1, N
+      OMPVV_TEST_AND_SET_VERBOSE(errors, anArray(i) .NE. (i + 10))
+      IF (errors .NE. 0) THEN
+        exit
+      END IF
+    END DO
+
+    DEALLOCATE(anArray)
+    unified_shared_memory_allocatable_map = errors
+  END FUNCTION unified_shared_memory_allocatable_map
+END PROGRAM test_requires_unified_shared_memory_allocatable_map

tests/5.0/requires/test_requires_unified_shared_memory_malloc.c

@@ -0,0 +1,84 @@
+//===---test_requires_unified_shared_memory_malloc.c --------------------------===//
+//
+// OpenMP API Version 5.0 Nov 2018
+// 
+// This test checks for unified shared memory of an array that is allocated using 
+// malloc on host is accessed from host and device with the same pointer.
+//
+// It uses the default mapping of pointers to access the array.
+//
+////===----------------------------------------------------------------------===//
+#include <omp.h>
+#include <stdio.h>
+#include <stdlib.h>
+#include "ompvv.h"
+
+#define N 1024
+
+#pragma omp requires unified_shared_memory
+
+int unified_shared_memory_malloc() {
+  OMPVV_INFOMSG("Unified shared memory testing - Malloced Array");
+  int errors = 0;
+
+  int *anArray;
+  int *anArrayCopy;
+
+  anArray = (int*)malloc(sizeof(int)*N);
+  anArrayCopy = (int*)malloc(sizeof(int)*N);
+
+   if( anArray == NULL ) {
+    OMPVV_ERROR("Memory for anArray was not allocated");
+    OMPVV_RETURN(1);
+  }
+
+  if( anArrayCopy == NULL ) {
+    OMPVV_ERROR("Memory for anArrayCopy was not allocated");
+    OMPVV_RETURN(1);
+  }
+
+  for (int i = 0; i < N; i++) {
+    anArray[i] = i;
+    anArrayCopy[i] = 0;
+  }
+  // Modify in the device
+#pragma omp target 
+  {
+    for (int i = 0; i < N; i++) {
+      anArray[i] += 10;
+    }
+  }
+  // Modify again on the host
+  for (int i = 0; i < N; i++) {
+    anArray[i] += 10;
+  }
+
+  // Get the value the device is seeing
+#pragma omp target 
+  {
+    for (int i = 0; i < N; i++) {
+      anArrayCopy[i] = anArray[i];
+    }
+  }
+
+  for (int i = 0; i < N; i++) {
+    OMPVV_TEST_AND_SET_VERBOSE(errors, anArray[i] != i + 20);
+    OMPVV_TEST_AND_SET_VERBOSE(errors, anArrayCopy[i] != i + 20);
+    if (errors) break;
+  }
+
+  free(anArray);
+  free(anArrayCopy);
+  return errors;
+}
+
+int main() {
+  int isOffloading;
+  OMPVV_TEST_AND_SET_OFFLOADING(isOffloading);
+  OMPVV_WARNING_IF(!isOffloading, "With no offloading, unified shared memory is guaranteed due to host execution");
+  int errors = 0;
+
+  OMPVV_TEST_AND_SET_VERBOSE(errors, unified_shared_memory_malloc());
+
+  OMPVV_REPORT_AND_RETURN(errors);
+}

tests/5.0/requires/test_requires_unified_shared_memory_malloc_is_device_ptr.c

@@ -0,0 +1,84 @@
+//===---test_requires_unified_shared_memory_malloc_is_device_ptr.c -----------===//
+//
+// OpenMP API Version 5.0 Nov 2018
+// 
+// This test Checks for unified shared memory of an array that is allocated using 
+// malloc and that is accessed from host and device with the same pointer
+//
+// To guarantee the use of the device pointer, we use the is_device_ptr clause
+//
+////===----------------------------------------------------------------------===//
+#include <omp.h>
+#include <stdio.h>
+#include <stdlib.h>
+#include "ompvv.h"
+
+#define N 1024
+
+#pragma omp requires unified_shared_memory
+
+int unified_shared_memory_malloc() {
+  OMPVV_INFOMSG("Unified shared memory testing - Array on malloc");
+  int errors = 0;
+
+  int *anArray;
+  int *anArrayCopy;
+
+  anArray = (int*)malloc(sizeof(int)*N);
+  anArrayCopy = (int*)malloc(sizeof(int)*N);
+
+  if( anArray == NULL ) {
+    OMPVV_ERROR("Memory for anArray was not allocated");
+    OMPVV_RETURN(1);
+  }
+
+  if( anArrayCopy == NULL ) {
+    OMPVV_ERROR("Memory for anArrayCopy was not allocated");
+    OMPVV_RETURN(1);
+  }
+
+  for (int i = 0; i < N; i++) {
+    anArray[i] = i;
+    anArrayCopy[i] = 0;
+  }
+  // Modify in the device
+#pragma omp target is_device_ptr(anArray)
+  {
+    for (int i = 0; i < N; i++) {
+      anArray[i] += 10;
+    }
+  }
+  // Modify again on the host
+  for (int i = 0; i < N; i++) {
+    anArray[i] += 10;
+  }
+
+  // Get the value the device is seeing
+#pragma omp target is_device_ptr(anArray)
+  {
+    for (int i = 0; i < N; i++) {
+      anArrayCopy[i] = anArray[i];
+    }
+  }
+
+  for (int i = 0; i < N; i++) {
+    OMPVV_TEST_AND_SET_VERBOSE(errors, anArray[i] != i + 20);
+    OMPVV_TEST_AND_SET_VERBOSE(errors, anArrayCopy[i] != i + 20);
+    if (errors) break;
+  }
+
+  free(anArray);
+  free(anArrayCopy);
+  return errors;
+}
+int main() {
+  int isOffloading;
+  OMPVV_TEST_AND_SET_OFFLOADING(isOffloading);
+  OMPVV_WARNING_IF(!isOffloading, "With no offloading, unified shared memory is guaranteed due to host execution");
+  int errors = 0;
+
+  OMPVV_TEST_AND_SET_VERBOSE(errors, unified_shared_memory_malloc());
+
+  OMPVV_REPORT_AND_RETURN(errors);
+}
+