diff --git a/clang/lib/Sema/SemaExprCXX.cpp b/clang/lib/Sema/SemaExprCXX.cpp index dc7ed4e9a48bc..3157119dd453e 100644 --- a/clang/lib/Sema/SemaExprCXX.cpp +++ b/clang/lib/Sema/SemaExprCXX.cpp @@ -852,7 +852,7 @@ ExprResult Sema::BuildCXXThrow(SourceLocation OpLoc, Expr *Ex, bool IsThrownVarInScope) { const llvm::Triple &T = Context.getTargetInfo().getTriple(); const bool IsOpenMPGPUTarget = - getLangOpts().OpenMPIsTargetDevice && (T.isNVPTX() || T.isAMDGCN()); + getLangOpts().OpenMPIsTargetDevice && T.isGPU(); DiagnoseExceptionUse(OpLoc, /* IsTry= */ false); diff --git a/clang/lib/Sema/SemaStmt.cpp b/clang/lib/Sema/SemaStmt.cpp index 5b3ef1adf38e3..655fa31bbf5c7 100644 --- a/clang/lib/Sema/SemaStmt.cpp +++ b/clang/lib/Sema/SemaStmt.cpp @@ -4357,7 +4357,7 @@ StmtResult Sema::ActOnCXXTryBlock(SourceLocation TryLoc, Stmt *TryBlock, ArrayRef Handlers) { const llvm::Triple &T = Context.getTargetInfo().getTriple(); const bool IsOpenMPGPUTarget = - getLangOpts().OpenMPIsTargetDevice && (T.isNVPTX() || T.isAMDGCN()); + getLangOpts().OpenMPIsTargetDevice && T.isGPU(); DiagnoseExceptionUse(TryLoc, /* IsTry= */ true); @@ -4464,7 +4464,7 @@ StmtResult Sema::ActOnCXXTryBlock(SourceLocation TryLoc, Stmt *TryBlock, void Sema::DiagnoseExceptionUse(SourceLocation Loc, bool IsTry) { const llvm::Triple &T = Context.getTargetInfo().getTriple(); const bool IsOpenMPGPUTarget = - getLangOpts().OpenMPIsTargetDevice && (T.isNVPTX() || T.isAMDGCN()); + getLangOpts().OpenMPIsTargetDevice && T.isGPU(); // Don't report an error if 'try' is used in system headers or in an OpenMP // target region compiled for a GPU architecture. diff --git a/clang/test/OpenMP/nvptx_target_exceptions_messages.cpp b/clang/test/OpenMP/target_exceptions_messages.cpp similarity index 54% rename from clang/test/OpenMP/nvptx_target_exceptions_messages.cpp rename to clang/test/OpenMP/target_exceptions_messages.cpp index 5d1d46cadff26..837fdb454c74f 100644 --- a/clang/test/OpenMP/nvptx_target_exceptions_messages.cpp +++ b/clang/test/OpenMP/target_exceptions_messages.cpp @@ -6,6 +6,22 @@ // RUN: -fopenmp-is-target-device -fopenmp-host-ir-file-path %t-ppc-host.bc -o - \ // RUN: -fexceptions -fcxx-exceptions -ferror-limit 100 +// RUN: %clang_cc1 -fopenmp -x c++ -triple powerpc64le-unknown-unknown \ +// RUN: -verify=host -fopenmp-targets=amdgcn-amd-amdhsa -emit-llvm-bc \ +// RUN: %s -o %t-ppc-host-amd.bc -fexceptions -fcxx-exceptions +// RUN: %clang_cc1 -verify -fopenmp -x c++ -triple amdgcn-amd-amdhsa \ +// RUN: -fopenmp-targets=amdgcn-amd-amdhsa -emit-llvm %s \ +// RUN: -fopenmp-is-target-device -fopenmp-host-ir-file-path %t-ppc-host-amd.bc -o - \ +// RUN: -fexceptions -fcxx-exceptions -ferror-limit 100 + +// RUN: %clang_cc1 -fopenmp -x c++ -triple powerpc64le-unknown-unknown \ +// RUN: -verify=host -fopenmp-targets=spirv64-intel -emit-llvm-bc \ +// RUN: %s -o %t-ppc-host-spirv.bc -fexceptions -fcxx-exceptions +// RUN: %clang_cc1 -verify -fopenmp -x c++ -triple spirv64-intel \ +// RUN: -fopenmp-targets=spirv64-intel -emit-llvm %s \ +// RUN: -fopenmp-is-target-device -fopenmp-host-ir-file-path %t-ppc-host-spirv.bc -o - \ +// RUN: -fexceptions -fcxx-exceptions -ferror-limit 100 + #ifndef HEADER #define HEADER @@ -34,7 +50,7 @@ T FA() { #pragma omp declare target struct S { int a; - S(int a) : a(a) { throw 1; } // expected-warning {{target 'nvptx64-unknown-unknown' does not support exception handling; 'throw' is assumed to be never reached}} + S(int a) : a(a) { throw 1; } // expected-warning-re {{target '{{.*}}' does not support exception handling; 'throw' is assumed to be never reached}} }; int foo() { return 0; } @@ -57,7 +73,7 @@ int maini1() { static long aaa = 23; a = foo() + bar() + b + c + d + aa + aaa + FA(); // expected-note{{called by 'maini1'}} if (!a) - throw "Error"; // expected-warning {{target 'nvptx64-unknown-unknown' does not support exception handling; 'throw' is assumed to be never reached}} + throw "Error"; // expected-warning-re {{target '{{.*}}' does not support exception handling; 'throw' is assumed to be never reached}} } } catch(...) { } @@ -67,14 +83,14 @@ int maini1() { int baz3() { return 2 + baz2(); } int baz2() { #pragma omp target - try { // expected-warning {{target 'nvptx64-unknown-unknown' does not support exception handling; 'catch' block is ignored}} + try { // expected-warning-re {{target '{{.*}}' does not support exception handling; 'catch' block is ignored}} ++c; } catch (...) { } return 2 + baz3(); } -int baz1() { throw 1; } // expected-warning {{target 'nvptx64-unknown-unknown' does not support exception handling; 'throw' is assumed to be never reached}} +int baz1() { throw 1; } // expected-warning-re {{target '{{.*}}' does not support exception handling; 'throw' is assumed to be never reached}} int foobar1(); int foobar2(); @@ -85,7 +101,7 @@ int (*B)() = &foobar2; #pragma omp end declare target int foobar1() { throw 1; } -int foobar2() { throw 1; } // expected-warning {{target 'nvptx64-unknown-unknown' does not support exception handling; 'throw' is assumed to be never reached}} +int foobar2() { throw 1; } // expected-warning-re {{target '{{.*}}' does not support exception handling; 'throw' is assumed to be never reached}} int foobar3(); @@ -95,7 +111,7 @@ int (*C)() = &foobar3; // expected-warning {{declaration is not declared in any int (*D)() = C; // expected-note {{used here}} // host-note@-1 {{used here}} #pragma omp end declare target -int foobar3() { throw 1; } // expected-warning {{target 'nvptx64-unknown-unknown' does not support exception handling; 'throw' is assumed to be never reached}} +int foobar3() { throw 1; } // expected-warning-re {{target '{{.*}}' does not support exception handling; 'throw' is assumed to be never reached}} // Check no infinite recursion in deferred diagnostic emitter. long E = (long)&E;