From b7ae08241c1bcd992959ea1246ae71bdd7dc6f06 Mon Sep 17 00:00:00 2001 From: srividya sundaram Date: Mon, 1 Nov 2021 20:33:07 -0700 Subject: [PATCH 1/4] [SYCL]Remove unwanted check for value dependency of const vars --- clang/include/clang/Sema/Sema.h | 3 +-- clang/lib/Sema/SemaExpr.cpp | 2 +- clang/lib/Sema/SemaSYCL.cpp | 4 ++-- 3 files changed, 4 insertions(+), 5 deletions(-) diff --git a/clang/include/clang/Sema/Sema.h b/clang/include/clang/Sema/Sema.h index 8e71c19ebc9d6..9ff90fc55c466 100644 --- a/clang/include/clang/Sema/Sema.h +++ b/clang/include/clang/Sema/Sema.h @@ -10593,8 +10593,7 @@ class Sema final { Expr *E); bool checkNSReturnsRetainedReturnType(SourceLocation loc, QualType type); - bool checkAllowedSYCLInitializer(VarDecl *VD, - bool CheckValueDependent = false); + bool checkAllowedSYCLInitializer(VarDecl *VD); //===--------------------------------------------------------------------===// // C++ Coroutines TS diff --git a/clang/lib/Sema/SemaExpr.cpp b/clang/lib/Sema/SemaExpr.cpp index 00840d32f1ca0..5fbbba68c8360 100644 --- a/clang/lib/Sema/SemaExpr.cpp +++ b/clang/lib/Sema/SemaExpr.cpp @@ -250,7 +250,7 @@ bool Sema::DiagnoseUseOfDecl(NamedDecl *D, ArrayRef Locs, // or constant-initialized. else if (IsRuntimeEvaluated && IsConst && VD->hasGlobalStorage() && !VD->isConstexpr() && - !checkAllowedSYCLInitializer(VD, /*CheckValueDependent =*/true)) + !checkAllowedSYCLInitializer(VD)) SYCLDiagIfDeviceCode(*Locs.begin(), diag::err_sycl_restrict) << Sema::KernelConstStaticVariable; } else if (auto *FDecl = dyn_cast(D)) { diff --git a/clang/lib/Sema/SemaSYCL.cpp b/clang/lib/Sema/SemaSYCL.cpp index fe2f405df22f6..eff50e3a3c649 100644 --- a/clang/lib/Sema/SemaSYCL.cpp +++ b/clang/lib/Sema/SemaSYCL.cpp @@ -4132,7 +4132,7 @@ void Sema::finalizeSYCLDelayedAnalysis(const FunctionDecl *Caller, } } -bool Sema::checkAllowedSYCLInitializer(VarDecl *VD, bool CheckValueDependent) { +bool Sema::checkAllowedSYCLInitializer(VarDecl *VD) { assert(getLangOpts().SYCLIsDevice && "Should only be called during SYCL compilation"); @@ -4140,7 +4140,7 @@ bool Sema::checkAllowedSYCLInitializer(VarDecl *VD, bool CheckValueDependent) { return true; const Expr *Init = VD->getInit(); - bool ValueDependent = CheckValueDependent && Init->isValueDependent(); + bool ValueDependent = Init->isValueDependent(); bool isConstantInit = Init && !ValueDependent && Init->isConstantInitializer(Context, false); if (!VD->isConstexpr() && Init && !ValueDependent && !isConstantInit) From 5e663dea64df9fab9cc3a6bf20b9f6592a618c82 Mon Sep 17 00:00:00 2001 From: srividya sundaram Date: Tue, 2 Nov 2021 06:15:36 -0700 Subject: [PATCH 2/4] clang-format fix --- clang/include/clang/Sema/Sema.h | 1 - clang/lib/Sema/SemaExpr.cpp | 3 +-- 2 files changed, 1 insertion(+), 3 deletions(-) diff --git a/clang/include/clang/Sema/Sema.h b/clang/include/clang/Sema/Sema.h index 9ff90fc55c466..df22cc6f1faa7 100644 --- a/clang/include/clang/Sema/Sema.h +++ b/clang/include/clang/Sema/Sema.h @@ -10594,7 +10594,6 @@ class Sema final { bool checkNSReturnsRetainedReturnType(SourceLocation loc, QualType type); bool checkAllowedSYCLInitializer(VarDecl *VD); - //===--------------------------------------------------------------------===// // C++ Coroutines TS // diff --git a/clang/lib/Sema/SemaExpr.cpp b/clang/lib/Sema/SemaExpr.cpp index 5fbbba68c8360..2b5335904b969 100644 --- a/clang/lib/Sema/SemaExpr.cpp +++ b/clang/lib/Sema/SemaExpr.cpp @@ -249,8 +249,7 @@ bool Sema::DiagnoseUseOfDecl(NamedDecl *D, ArrayRef Locs, // Disallow const statics and globals that are not zero-initialized // or constant-initialized. else if (IsRuntimeEvaluated && IsConst && VD->hasGlobalStorage() && - !VD->isConstexpr() && - !checkAllowedSYCLInitializer(VD)) + !VD->isConstexpr() && !checkAllowedSYCLInitializer(VD)) SYCLDiagIfDeviceCode(*Locs.begin(), diag::err_sycl_restrict) << Sema::KernelConstStaticVariable; } else if (auto *FDecl = dyn_cast(D)) { From ecd266c5ec01630e5422ac1f1da653995904a072 Mon Sep 17 00:00:00 2001 From: srividya sundaram Date: Tue, 9 Nov 2021 15:40:03 -0800 Subject: [PATCH 3/4] [SYCL]Add test --- clang/test/SemaSYCL/Inputs/sycl.hpp | 9 ++++++- .../spec-const-value-dependent-crash.cpp | 26 +++++++++++++++++++ 2 files changed, 34 insertions(+), 1 deletion(-) create mode 100644 clang/test/SemaSYCL/spec-const-value-dependent-crash.cpp diff --git a/clang/test/SemaSYCL/Inputs/sycl.hpp b/clang/test/SemaSYCL/Inputs/sycl.hpp index 90c5f3042cb66..9cb22bce7177b 100644 --- a/clang/test/SemaSYCL/Inputs/sycl.hpp +++ b/clang/test/SemaSYCL/Inputs/sycl.hpp @@ -319,7 +319,14 @@ namespace ext { namespace oneapi { namespace experimental { template -class spec_constant {}; +class spec_constant { +public: + spec_constant() {} + explicit constexpr spec_constant(T defaultVal) : DefaultValue(defaultVal) {} + +private: + T DefaultValue; +}; } // namespace experimental } // namespace oneapi } // namespace ext diff --git a/clang/test/SemaSYCL/spec-const-value-dependent-crash.cpp b/clang/test/SemaSYCL/spec-const-value-dependent-crash.cpp new file mode 100644 index 0000000000000..ef1f1c66a73d3 --- /dev/null +++ b/clang/test/SemaSYCL/spec-const-value-dependent-crash.cpp @@ -0,0 +1,26 @@ +// RUN: %clang_cc1 -fsycl-is-device -internal-isystem %S/Inputs -ast-dump -sycl-std=2020 %s | FileCheck %s +// This test checks that Clang doesn't crash if a specialization constant is +// value dependent. + + + +#include "sycl.hpp" +sycl::queue myQueue; + +int main() { + constexpr int default_val = 20; + cl::sycl::ext::oneapi::experimental::spec_constant SC(default_val); + + myQueue.submit([&](sycl::handler &h) { + h.single_task( + [=] { + cl::sycl::ext::oneapi::experimental::spec_constant res = SC; + }); + }); + return 0; +} + +// CHECK: FunctionDecl {{.*}}kernel_sc{{.*}} 'void ()' +// CHECK: VarDecl {{.*}}'(lambda at {{.*}}' +// CHECK-NEXT: InitListExpr {{.*}}'(lambda at {{.*}}' +// CHECK-NEXT: CXXConstructExpr {{.*}}'cl::sycl::ext::oneapi::experimental::spec_constant':'sycl::ext::oneapi::experimental::spec_constant' 'void ()' \ No newline at end of file From d7168110b105e1292629ed5e81d884618bed12bc Mon Sep 17 00:00:00 2001 From: Srividya Sundaram Date: Tue, 9 Nov 2021 19:02:24 -0800 Subject: [PATCH 4/4] Update spec-const-value-dependent-crash.cpp --- clang/test/SemaSYCL/spec-const-value-dependent-crash.cpp | 4 +--- 1 file changed, 1 insertion(+), 3 deletions(-) diff --git a/clang/test/SemaSYCL/spec-const-value-dependent-crash.cpp b/clang/test/SemaSYCL/spec-const-value-dependent-crash.cpp index ef1f1c66a73d3..7ddb5671add1f 100644 --- a/clang/test/SemaSYCL/spec-const-value-dependent-crash.cpp +++ b/clang/test/SemaSYCL/spec-const-value-dependent-crash.cpp @@ -2,8 +2,6 @@ // This test checks that Clang doesn't crash if a specialization constant is // value dependent. - - #include "sycl.hpp" sycl::queue myQueue; @@ -23,4 +21,4 @@ int main() { // CHECK: FunctionDecl {{.*}}kernel_sc{{.*}} 'void ()' // CHECK: VarDecl {{.*}}'(lambda at {{.*}}' // CHECK-NEXT: InitListExpr {{.*}}'(lambda at {{.*}}' -// CHECK-NEXT: CXXConstructExpr {{.*}}'cl::sycl::ext::oneapi::experimental::spec_constant':'sycl::ext::oneapi::experimental::spec_constant' 'void ()' \ No newline at end of file +// CHECK-NEXT: CXXConstructExpr {{.*}}'cl::sycl::ext::oneapi::experimental::spec_constant':'sycl::ext::oneapi::experimental::spec_constant' 'void ()'