https://github.com/sarnex created https://github.com/llvm/llvm-project/pull/169094
More missed target checks. >From 68d3acfd1ffdeada907962f040101a75d261ff7c Mon Sep 17 00:00:00 2001 From: Nick Sarnie <[email protected]> Date: Fri, 21 Nov 2025 13:09:50 -0800 Subject: [PATCH] [OpenMP][SPIRV] Disable exceptions for OpenMP SPIR-V Signed-off-by: Nick Sarnie <[email protected]> --- clang/lib/CodeGen/CGException.cpp | 4 ++-- clang/lib/Frontend/CompilerInvocation.cpp | 3 +-- clang/test/OpenMP/spirv_target_codegen_noexceptions.cpp | 9 +++++++++ 3 files changed, 12 insertions(+), 4 deletions(-) create mode 100644 clang/test/OpenMP/spirv_target_codegen_noexceptions.cpp diff --git a/clang/lib/CodeGen/CGException.cpp b/clang/lib/CodeGen/CGException.cpp index f86af4581c345..e9d20672ce185 100644 --- a/clang/lib/CodeGen/CGException.cpp +++ b/clang/lib/CodeGen/CGException.cpp @@ -450,7 +450,7 @@ void CodeGenFunction::EmitCXXThrowExpr(const CXXThrowExpr *E, // Therefore, we emit a trap which will abort the program, and // prompt a warning indicating that a trap will be emitted. const llvm::Triple &T = Target.getTriple(); - if (CGM.getLangOpts().OpenMPIsTargetDevice && (T.isNVPTX() || T.isAMDGCN())) { + if (CGM.getLangOpts().OpenMPIsTargetDevice && T.isGPU()) { EmitTrapCall(llvm::Intrinsic::trap); return; } @@ -627,7 +627,7 @@ void CodeGenFunction::EmitCXXTryStmt(const CXXTryStmt &S) { // If we encounter a try statement on in an OpenMP target region offloaded to // a GPU, we treat it as a basic block. const bool IsTargetDevice = - (CGM.getLangOpts().OpenMPIsTargetDevice && (T.isNVPTX() || T.isAMDGCN())); + (CGM.getLangOpts().OpenMPIsTargetDevice && T.isGPU()); if (!IsTargetDevice) EnterCXXTryStmt(S); EmitStmt(S.getTryBlock()); diff --git a/clang/lib/Frontend/CompilerInvocation.cpp b/clang/lib/Frontend/CompilerInvocation.cpp index 2e027934a8db6..15b44ef1df3da 100644 --- a/clang/lib/Frontend/CompilerInvocation.cpp +++ b/clang/lib/Frontend/CompilerInvocation.cpp @@ -4303,8 +4303,7 @@ bool CompilerInvocation::ParseLangArgs(LangOptions &Opts, ArgList &Args, // Set the flag to prevent the implementation from emitting device exception // handling code for those requiring so. - if ((Opts.OpenMPIsTargetDevice && (T.isNVPTX() || T.isAMDGCN())) || - Opts.OpenCLCPlusPlus) { + if ((Opts.OpenMPIsTargetDevice && T.isGPU()) || Opts.OpenCLCPlusPlus) { Opts.Exceptions = 0; Opts.CXXExceptions = 0; diff --git a/clang/test/OpenMP/spirv_target_codegen_noexceptions.cpp b/clang/test/OpenMP/spirv_target_codegen_noexceptions.cpp new file mode 100644 index 0000000000000..42f8f3ea70f7d --- /dev/null +++ b/clang/test/OpenMP/spirv_target_codegen_noexceptions.cpp @@ -0,0 +1,9 @@ +// RUN: %clang_cc1 -fexceptions -fcxx-exceptions -Wno-openmp-target-exception -fopenmp -x c++ -triple x86_64-unknown-linux -fopenmp-targets=spirv64-intel -emit-llvm-bc %s -o %t-host.bc +// RUN: %clang_cc1 -fexceptions -fcxx-exceptions -Wno-openmp-target-exception -fopenmp -x c++ -triple spirv64-intel -fopenmp-targets=spirv64-intel -emit-llvm %s -fopenmp-is-target-device -fopenmp-host-ir-file-path %t-host.bc -o - | \ +// RUN: FileCheck -implicit-check-not='{{invoke|throw|cxa}}' %s +void foo() { + // CHECK: call addrspace(9) void @llvm.trap() + // CHECK-NEXT: call spir_func addrspace(9) void @__kmpc_target_deinit() + #pragma omp target + throw "bad"; +} _______________________________________________ cfe-commits mailing list [email protected] https://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits
