[Openmp-commits] [PATCH] D108255: [Libomptarget] Correctly default to Generic if exec_mode is not present

[Joseph Huber via Phabricator via Openmp-commits]

jhuber6 created this revision.
jhuber6 added a reviewer: jdoerfert.
jhuber6 requested review of this revision.
Herald added a project: OpenMP.
Herald added a subscriber: openmp-commits.

Currently, the runtime returns an error when the `exec_mode` global is
not present. The expected behvaiour is that the region will default to
Generic. This prevents global constructors from being called because
they do not contain execution mode globals.


Repository:
  rG LLVM Github Monorepo

https://reviews.llvm.org/D108255

Files:
  openmp/libomptarget/plugins/cuda/src/rtl.cpp
  openmp/libomptarget/test/offloading/global_constructor.cpp


Index: openmp/libomptarget/test/offloading/global_constructor.cpp
===================================================================
--- /dev/null
+++ openmp/libomptarget/test/offloading/global_constructor.cpp
@@ -0,0 +1,19 @@
+// RUN: %libomptarget-compilexx-generic && %libomptarget-run-generic | %fcheck-generic
+
+#include <cmath>
+#include <cstdio>
+
+const double Host = log(2.0) / log(2.0);
+#pragma omp declare target
+const double Device = log(2.0) / log(2.0);
+#pragma omp end declare target
+
+int main() {
+  double X;
+#pragma omp target map(from : X)
+  { X = Device; }
+
+  // CHECK: PASS
+  if (X == Host)
+    printf("PASS\n");
+}
Index: openmp/libomptarget/plugins/cuda/src/rtl.cpp
===================================================================
--- openmp/libomptarget/plugins/cuda/src/rtl.cpp
+++ openmp/libomptarget/plugins/cuda/src/rtl.cpp
@@ -829,10 +829,9 @@
           return nullptr;
         }
       } else {
-        REPORT("Loading global exec_mode '%s' - symbol missing, using default "
-               "value GENERIC (1)\n",
-               ExecModeName);
-        CUDA_ERR_STRING(Err);
+        DP("Loading global exec_mode '%s' - symbol missing, using default "
+           "value GENERIC (1)\n",
+           ExecModeName);
       }
 
       KernelsList.emplace_back(Func, ExecModeVal);


-------------- next part --------------
A non-text attachment was scrubbed...
Name: D108255.367059.patch
Type: text/x-patch
Size: 1326 bytes
Desc: not available
URL: <http://lists.llvm.org/pipermail/openmp-commits/attachments/20210817/a898a044/attachment.bin>