[OpenMP] target_data_begin: fail on device alloc fail

[jdenny]

Summary

Without this patch, target_data_begin continues after an illegal
mapping or an out-of-memory error on the device. With this patch, it
terminates the runtime with an error instead.

The new test exercises only illegal mappings. I didn't think of a
good way to exercise out-of-memory errors from the test suite.

[jdenny]

Extracted test infrastructure changes to D78566.

Rebased.

[jdoerfert]

I assume this works also if we are not on those platforms.

LGTM.

[jdenny]

In D78170#1994836, @jdoerfert wrote:

I assume this works also if we are not on those platforms.

It works on the platforms I tried: it runs the commands for the current platform and effectively skips (that is, calls echo instead) commands for the others.

LGTM.

Thanks.

openmp/libomptarget/src/omptarget.cpp

for (int32_t i = 0; i < arg_num; ++i) {
 
     void *TgtPtrBegin = Device.getOrAllocTgtPtr(HstPtrBegin, HstPtrBase,
         data_size, IsNew, IsHostPtr, IsImplicit, UpdateRef, HasCloseModifier);
     if (!TgtPtrBegin && data_size) {
       // If data_size==0, then the argument could be a zero-length pointer to
       // NULL, so getOrAlloc() returning NULL is not an error.
       DP("Call to getOrAllocTgtPtr returned null pointer (device failure or "
           "illegal mapping).\n");
+      return OFFLOAD_FAIL;
     }
     DP("There are %" PRId64 " bytes allocated at target address " DPxMOD
         " - is%s new\n", data_size, DPxPTR(TgtPtrBegin),
         (IsNew ? "" : " not"));
 
     if (arg_types[i] & OMP_TGT_MAPTYPE_RETURN_PARAM) {
       uintptr_t Delta = (uintptr_t)HstPtrBegin - (uintptr_t)HstPtrBase;
       void *TgtPtrBase = (void *)((uintptr_t)TgtPtrBegin - Delta);

openmp/libomptarget/test/mapping/alloc_fail.c

+// RUN: %libomptarget-compile-aarch64-unknown-linux-gnu
+// RUN: %libomptarget-run-fail-aarch64-unknown-linux-gnu 2>&1 \
+// RUN: | %fcheck-aarch64-unknown-linux-gnu
+
+// RUN: %libomptarget-compile-powerpc64-ibm-linux-gnu
+// RUN: %libomptarget-run-fail-powerpc64-ibm-linux-gnu 2>&1 \
+// RUN: | %fcheck-powerpc64-ibm-linux-gnu
+
+// RUN: %libomptarget-compile-powerpc64le-ibm-linux-gnu
+// 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: %libomptarget-run-fail-x86_64-pc-linux-gnu 2>&1 \
+// RUN: | %fcheck-x86_64-pc-linux-gnu
+
+// CHECK: Libomptarget fatal error 1: failure of target construct while offloading is mandatory
+
+int main() {
+  int arr[4] = {0, 1, 2, 3};
+#pragma omp target data map(alloc: arr[0:2])
+#pragma omp target data map(alloc: arr[1:2])
+  ;
+  return 0;
+}