Index: openmp/libomptarget/src/device.cpp =================================================================== --- openmp/libomptarget/src/device.cpp +++ openmp/libomptarget/src/device.cpp @@ -192,6 +192,10 @@ } else if ((lr.Flags.ExtendsBefore || lr.Flags.ExtendsAfter) && !IsImplicit) { // Explicit extension of mapped data - not allowed. DP("Explicit extension of mapping is not allowed.\n"); + if (HasPresentModifier) + MESSAGE("device mapping required by 'present' map type modifier does not " + "exist for host address " DPxMOD " (%ld bytes)", + DPxPTR(HstPtrBegin), Size); } else if (RTLs->RequiresFlags & OMP_REQ_UNIFIED_SHARED_MEMORY && !HasCloseModifier) { // If unified shared memory is active, implicitly mapped variables that are Index: openmp/libomptarget/test/mapping/present/target_array_extension.c =================================================================== --- /dev/null +++ openmp/libomptarget/test/mapping/present/target_array_extension.c @@ -0,0 +1,94 @@ +// -------------------------------------------------- +// Check extends before +// -------------------------------------------------- + +// RUN: %libomptarget-compile-aarch64-unknown-linux-gnu \ +// RUN: -fopenmp-version=51 -DEXTENDS=BEFORE +// RUN: %libomptarget-run-fail-aarch64-unknown-linux-gnu 2>&1 \ +// RUN: | %fcheck-aarch64-unknown-linux-gnu + +// RUN: %libomptarget-compile-powerpc64-ibm-linux-gnu \ +// RUN: -fopenmp-version=51 -DEXTENDS=BEFORE +// RUN: %libomptarget-run-fail-powerpc64-ibm-linux-gnu 2>&1 \ +// RUN: | %fcheck-powerpc64-ibm-linux-gnu + +// RUN: %libomptarget-compile-powerpc64le-ibm-linux-gnu \ +// RUN: -fopenmp-version=51 -DEXTENDS=BEFORE +// RUN: %libomptarget-run-fail-powerpc64le-ibm-linux-gnu 2>&1 \ +// RUN: | %fcheck-powerpc64le-ibm-linux-gnu + +// RUN: %libomptarget-compile-x86_64-pc-linux-gnu \ +// RUN: -fopenmp-version=51 -DEXTENDS=BEFORE +// RUN: %libomptarget-run-fail-x86_64-pc-linux-gnu 2>&1 \ +// RUN: | %fcheck-x86_64-pc-linux-gnu + +// -------------------------------------------------- +// Check extends after +// -------------------------------------------------- + +// RUN: %libomptarget-compile-aarch64-unknown-linux-gnu \ +// RUN: -fopenmp-version=51 -DEXTENDS=AFTER +// RUN: %libomptarget-run-fail-aarch64-unknown-linux-gnu 2>&1 \ +// RUN: | %fcheck-aarch64-unknown-linux-gnu + +// RUN: %libomptarget-compile-powerpc64-ibm-linux-gnu \ +// RUN: -fopenmp-version=51 -DEXTENDS=AFTER +// RUN: %libomptarget-run-fail-powerpc64-ibm-linux-gnu 2>&1 \ +// RUN: | %fcheck-powerpc64-ibm-linux-gnu + +// RUN: %libomptarget-compile-powerpc64le-ibm-linux-gnu \ +// RUN: -fopenmp-version=51 -DEXTENDS=AFTER +// RUN: %libomptarget-run-fail-powerpc64le-ibm-linux-gnu 2>&1 \ +// RUN: | %fcheck-powerpc64le-ibm-linux-gnu + +// RUN: %libomptarget-compile-x86_64-pc-linux-gnu \ +// RUN: -fopenmp-version=51 -DEXTENDS=AFTER +// RUN: %libomptarget-run-fail-x86_64-pc-linux-gnu 2>&1 \ +// RUN: | %fcheck-x86_64-pc-linux-gnu + +// END. + +#include + +#define BEFORE 0 +#define AFTER 1 + +#if EXTENDS == BEFORE +# define SMALL 98:2 +# define LARGE 0:100 +#elif EXTENDS == AFTER +# define SMALL 0:2 +# define LARGE 0:100 +#else +# error EXTENDS undefined +#endif + +int main() { + int arr[100]; + + // CHECK: addr=0x[[#%x,HOST_ADDR:]], size=[[#%u,SIZE:]] + fprintf(stderr, "addr=%p, size=%ld\n", arr, sizeof arr); + + // CHECK-NOT: Libomptarget +#pragma omp target data map(alloc: arr[LARGE]) + { +#pragma omp target map(present, tofrom: arr[SMALL]) + ; + } + + // CHECK: arr is present + fprintf(stderr, "arr is present\n"); + + // CHECK: Libomptarget message: device mapping required by 'present' map type modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] ([[#SIZE]] bytes) + // CHECK: Libomptarget fatal error 1: failure of target construct while offloading is mandatory +#pragma omp target data map(alloc: arr[SMALL]) + { +#pragma omp target map(present, tofrom: arr[LARGE]) + ; + } + + // CHECK-NOT: arr is present + fprintf(stderr, "arr is present\n"); + + return 0; +} Index: openmp/libomptarget/test/mapping/present/target_data_array_extension.c =================================================================== --- /dev/null +++ openmp/libomptarget/test/mapping/present/target_data_array_extension.c @@ -0,0 +1,94 @@ +// -------------------------------------------------- +// Check extends before +// -------------------------------------------------- + +// RUN: %libomptarget-compile-aarch64-unknown-linux-gnu \ +// RUN: -fopenmp-version=51 -DEXTENDS=BEFORE +// RUN: %libomptarget-run-fail-aarch64-unknown-linux-gnu 2>&1 \ +// RUN: | %fcheck-aarch64-unknown-linux-gnu + +// RUN: %libomptarget-compile-powerpc64-ibm-linux-gnu \ +// RUN: -fopenmp-version=51 -DEXTENDS=BEFORE +// RUN: %libomptarget-run-fail-powerpc64-ibm-linux-gnu 2>&1 \ +// RUN: | %fcheck-powerpc64-ibm-linux-gnu + +// RUN: %libomptarget-compile-powerpc64le-ibm-linux-gnu \ +// RUN: -fopenmp-version=51 -DEXTENDS=BEFORE +// RUN: %libomptarget-run-fail-powerpc64le-ibm-linux-gnu 2>&1 \ +// RUN: | %fcheck-powerpc64le-ibm-linux-gnu + +// RUN: %libomptarget-compile-x86_64-pc-linux-gnu \ +// RUN: -fopenmp-version=51 -DEXTENDS=BEFORE +// RUN: %libomptarget-run-fail-x86_64-pc-linux-gnu 2>&1 \ +// RUN: | %fcheck-x86_64-pc-linux-gnu + +// -------------------------------------------------- +// Check extends after +// -------------------------------------------------- + +// RUN: %libomptarget-compile-aarch64-unknown-linux-gnu \ +// RUN: -fopenmp-version=51 -DEXTENDS=AFTER +// RUN: %libomptarget-run-fail-aarch64-unknown-linux-gnu 2>&1 \ +// RUN: | %fcheck-aarch64-unknown-linux-gnu + +// RUN: %libomptarget-compile-powerpc64-ibm-linux-gnu \ +// RUN: -fopenmp-version=51 -DEXTENDS=AFTER +// RUN: %libomptarget-run-fail-powerpc64-ibm-linux-gnu 2>&1 \ +// RUN: | %fcheck-powerpc64-ibm-linux-gnu + +// RUN: %libomptarget-compile-powerpc64le-ibm-linux-gnu \ +// RUN: -fopenmp-version=51 -DEXTENDS=AFTER +// RUN: %libomptarget-run-fail-powerpc64le-ibm-linux-gnu 2>&1 \ +// RUN: | %fcheck-powerpc64le-ibm-linux-gnu + +// RUN: %libomptarget-compile-x86_64-pc-linux-gnu \ +// RUN: -fopenmp-version=51 -DEXTENDS=AFTER +// RUN: %libomptarget-run-fail-x86_64-pc-linux-gnu 2>&1 \ +// RUN: | %fcheck-x86_64-pc-linux-gnu + +// END. + +#include + +#define BEFORE 0 +#define AFTER 1 + +#if EXTENDS == BEFORE +# define SMALL 98:2 +# define LARGE 0:100 +#elif EXTENDS == AFTER +# define SMALL 0:2 +# define LARGE 0:100 +#else +# error EXTENDS undefined +#endif + +int main() { + int arr[100]; + + // CHECK: addr=0x[[#%x,HOST_ADDR:]], size=[[#%u,SIZE:]] + fprintf(stderr, "addr=%p, size=%ld\n", arr, sizeof arr); + + // CHECK-NOT: Libomptarget +#pragma omp target data map(alloc: arr[LARGE]) + { +#pragma omp target data map(present, tofrom: arr[SMALL]) + ; + } + + // CHECK: arr is present + fprintf(stderr, "arr is present\n"); + + // CHECK: Libomptarget message: device mapping required by 'present' map type modifier does not exist for host address 0x{{0*}}[[#HOST_ADDR]] ([[#SIZE]] bytes) + // CHECK: Libomptarget fatal error 1: failure of target construct while offloading is mandatory +#pragma omp target data map(alloc: arr[SMALL]) + { +#pragma omp target data map(present, tofrom: arr[LARGE]) + ; + } + + // CHECK-NOT: arr is present + fprintf(stderr, "arr is present\n"); + + return 0; +}