[PATCH] D29904: [OpenMP] Prevent emission of exception handling code when using OpenMP to offload to NVIDIA devices.

[Gheorghe-Teodor Bercea via Phabricator via cfe-commits]

gtbercea created this revision.

For the OpenMP toolchain which offloads to NVIDIA GPUs make sure that no exception handling code is emitted.


Repository:
  rL LLVM

https://reviews.llvm.org/D29904

Files:
  lib/Frontend/CompilerInvocation.cpp
  test/OpenMP/target_parallel_no_exceptions.cpp


Index: test/OpenMP/target_parallel_no_exceptions.cpp
===================================================================
--- /dev/null
+++ test/OpenMP/target_parallel_no_exceptions.cpp
@@ -0,0 +1,29 @@
+/// Make sure no exception messages are inclided in the llvm output.
+// RUN: %clang -fopenmp -S -emit-llvm -fopenmp -fopenmp-targets=nvptx64-nvidia-cuda %s -o - 2>&1 | FileCheck -check-prefix=CHK-EXCEPTION %s
+
+#include <stdio.h>
+
+#define SIZE 1000000
+#define EPS 1e-10
+
+void test_increment(){
+  #pragma omp target
+  {
+    #pragma omp parallel
+    {
+      int a = 10;
+      printf("a = %d\n", a);
+    }
+  }
+}
+
+int main(){
+  test_increment();
+  return 0;
+}
+
+//CHK-EXCEPTION: __CLANG_OFFLOAD_BUNDLE____START__ openmp-nvptx64-nvidia-cuda
+//CHK-EXCEPTION-NOT: __cxa_begin_catch
+//CHK-EXCEPTION-NOT: terminate.lpad
+//CHK-EXCEPTION: __CLANG_OFFLOAD_BUNDLE____END__ openmp-nvptx64-nvidia-cuda
+
Index: lib/Frontend/CompilerInvocation.cpp
===================================================================
--- lib/Frontend/CompilerInvocation.cpp
+++ lib/Frontend/CompilerInvocation.cpp
@@ -2163,6 +2163,13 @@
         break;
       }
     }
+
+    // Set the flag to prevent the implementation from emitting device exception
+    // handling code for those requiring so.
+    if (Opts.OpenMPIsDevice && T.isNVPTX()) {
+      Opts.Exceptions = 0;
+      Opts.CXXExceptions = 0;
+    }
   }
 
   // Get the OpenMP target triples if any.


-------------- next part --------------
A non-text attachment was scrubbed...
Name: D29904.88241.patch
Type: text/x-patch
Size: 1457 bytes
Desc: not available
URL: <http://lists.llvm.org/pipermail/cfe-commits/attachments/20170213/e3808a5c/attachment.bin>