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. 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 + +#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 +