Remove spec_constant support from clang

Author: AlexeySachkov

clang/include/clang/Basic/Attr.td

@@ -1354,12 +1354,12 @@ def SYCLType: InheritableAttr {
   let Subjects = SubjectList<[CXXRecord, Enum], ErrorDiag>;
   let LangOpts = [SYCLIsDevice, SilentlyIgnoreSYCLIsHost];
   let Args = [EnumArgument<"Type", "SYCLType",
-                           ["accessor", "local_accessor", "spec_constant",
+                           ["accessor", "local_accessor",
                             "specialization_id", "kernel_handler", "buffer_location",
                             "no_alias", "accessor_property_list", "group",
                             "private_memory", "aspect", "annotated_ptr", "annotated_arg",
                             "stream", "sampler", "host_pipe"],
-                           ["accessor", "local_accessor", "spec_constant",
+                           ["accessor", "local_accessor",
                             "specialization_id", "kernel_handler", "buffer_location",
                             "no_alias", "accessor_property_list", "group",
                             "private_memory", "aspect", "annotated_ptr", "annotated_arg",

clang/lib/Sema/SemaSYCL.cpp

@@ -1257,8 +1257,6 @@ class KernelObjVisitor {
                   QualType FieldTy, HandlerTys &... Handlers) {
     if (isSyclSpecialType(FieldTy, SemaRef))
       KF_FOR_EACH(handleSyclSpecialType, Field, FieldTy);
-    else if (isSyclType(FieldTy, SYCLTypeAttr::spec_constant))
-      KF_FOR_EACH(handleSyclSpecConstantType, Field, FieldTy);
     else if (FieldTy->isStructureOrClassType()) {
       if (KF_FOR_EACH(handleStructType, Field, FieldTy)) {
         CXXRecordDecl *RD = FieldTy->getAsCXXRecordDecl();
@@ -1327,10 +1325,6 @@ class SyclKernelFieldHandlerBase {
   }
   virtual bool handleSyclSpecialType(FieldDecl *, QualType) { return true; }
 
-  virtual bool handleSyclSpecConstantType(FieldDecl *, QualType) {
-    return true;
-  }
-
   virtual bool handleStructType(FieldDecl *, QualType) { return true; }
   virtual bool handleUnionType(FieldDecl *, QualType) { return true; }
   virtual bool handleReferenceType(FieldDecl *, QualType) { return true; }
@@ -1826,11 +1820,6 @@ class SyclKernelDecompMarker : public SyclKernelFieldHandler {
     return true;
   }
 
-  bool handleSyclSpecConstantType(FieldDecl *, QualType) final {
-    CollectionStack.back() = true;
-    return true;
-  }
-
   bool handlePointerType(FieldDecl *, QualType) final {
     PointerStack.back() = true;
     return true;
@@ -3448,10 +3437,6 @@ class SyclKernelBodyCreator : public SyclKernelFieldHandler {
     return handleSpecialType(BS, Ty);
   }
 
-  bool handleSyclSpecConstantType(FieldDecl *FD, QualType Ty) final {
-    return handleSpecialType(FD, Ty);
-  }
-
   bool handlePointerType(FieldDecl *FD, QualType FieldTy) final {
     Expr *PointerRef =
         createPointerParamReferenceExpr(FieldTy, StructDepth != 0);
@@ -3717,21 +3702,6 @@ class SyclKernelIntHeaderCreator : public SyclKernelFieldHandler {
     return true;
   }
 
-  bool handleSyclSpecConstantType(FieldDecl *FD, QualType FieldTy) final {
-    const TemplateArgumentList &TemplateArgs =
-        cast<ClassTemplateSpecializationDecl>(FieldTy->getAsRecordDecl())
-            ->getTemplateInstantiationArgs();
-    assert(TemplateArgs.size() == 2 &&
-           "Incorrect template args for spec constant type");
-    // Get specialization constant ID type, which is the second template
-    // argument.
-    QualType SpecConstIDTy = TemplateArgs.get(1).getAsType().getCanonicalType();
-    const std::string SpecConstName = SYCLUniqueStableNameExpr::ComputeName(
-        SemaRef.getASTContext(), SpecConstIDTy);
-    Header.addSpecConstant(SpecConstName, SpecConstIDTy);
-    return true;
-  }
-
   bool handlePointerType(FieldDecl *FD, QualType FieldTy) final {
     addParam(FD, FieldTy,
              ((StructDepth) ? SYCLIntegrationHeader::kind_std_layout

clang/test/CodeGenSYCL/Inputs/sycl.hpp

@@ -370,20 +370,6 @@ struct get_kernel_name_t<auto_name, Type> {
 namespace ext {
 namespace oneapi {
 namespace experimental {
-template <typename T, typename ID = T>
-class __SYCL_TYPE(spec_constant) spec_constant {
-public:
-  spec_constant() {}
-  spec_constant(T Cst) {}
-
-  T get() const { // explicit access.
-    return T();   // Dummy implementaion.
-  }
-  operator T() const { // implicit conversion.
-    return get();
-  }
-};
-
 #ifdef __SYCL_DEVICE_ONLY__
 #define __SYCL_CONSTANT_AS __attribute__((opencl_constant))
 #else

clang/test/CodeGenSYCL/int_header_spec_const.cpp

@@ -344,20 +344,6 @@ struct get_kernel_name_t<auto_name, Type> {
 namespace ext {
 namespace oneapi {
 namespace experimental {
-template <typename T, typename ID = T>
-class __SYCL_TYPE(spec_constant) spec_constant {
-public:
-  spec_constant() {}
-  spec_constant(T Cst) {}
-
-  T get() const { // explicit access.
-    return T();   // Dummy implementaion.
-  }
-  operator T() const { // implicit conversion.
-    return get();
-  }
-};
-
 #ifdef __SYCL_DEVICE_ONLY__
 #define __SYCL_CONSTANT_AS __attribute__((opencl_constant))
 #else

clang/test/Frontend/Inputs/sycl.hpp

@@ -395,16 +395,6 @@ template <typename T, access::address_space AS> class multi_ptr {
 namespace ext {
 namespace oneapi {
 namespace experimental {
-template <typename T, typename ID = T>
-class __SYCL_TYPE(spec_constant) spec_constant {
-public:
-  spec_constant() {}
-  explicit constexpr spec_constant(T defaultVal) : DefaultValue(defaultVal) {}
-
-private:
-  T DefaultValue;
-};
-
 template <typename T, typename... Props>
 class __attribute__((sycl_special_class)) __SYCL_TYPE(annotated_arg) annotated_arg {
   T obj;

clang/test/SemaSYCL/Inputs/sycl.hpp

@@ -21,10 +21,6 @@ struct StructWithSampler {
   sycl::sampler sampl;
 };
 
-struct StructWithSpecConst {
-  sycl::ext::oneapi::experimental::spec_constant<int, class f1> SC;
-};
-
 sycl::handler H;
 
 struct StructWithStream {
@@ -136,20 +132,6 @@ int main() {
     // CHECK: FunctionDecl {{.*}}Sampl2{{.*}} 'void (sampler_t, StructNonDecomposed, int)'
   }
 
-  {
-    StructWithArray<StructWithSpecConst> t1;
-    myQueue.submit([&](sycl::handler &h) {
-      h.single_task<class SpecConst1>([=]() { return t1.i; });
-    });
-    // CHECK: FunctionDecl {{.*}}SpecConst{{.*}} 'void (StructNonDecomposed, int)'
-
-    DerivedStruct<StructWithSpecConst> t2;
-    myQueue.submit([&](sycl::handler &h) {
-      h.single_task<class SpecConst2>([=]() { return t2.i; });
-    });
-    // CHECK: FunctionDecl {{.*}}SpecConst2{{.*}} 'void (StructNonDecomposed, int)'
-  }
-
   {
     StructWithArray<StructWithStream> t1;
     myQueue.submit([&](sycl::handler &h) {

clang/test/SemaSYCL/decomposition.cpp

@@ -20,12 +20,6 @@ class [[__sycl_detail__::sycl_type(local_accessor)]] local_accessor {};
 // CHECK: CXXRecordDecl {{.*}} class local_accessor definition
 // CHECK: SYCLTypeAttr {{.*}} local_accessor
 
-class [[__sycl_detail__::sycl_type(spec_constant)]]
-[[__sycl_detail__::sycl_type(spec_constant)]] spec_constant;
-// CHECK: CXXRecordDecl {{.*}} spec_constant
-// CHECK: SYCLTypeAttr {{.*}} spec_constant
-// CHECK-NOT: SYCLTypeAttr {{.*}} spec_constant
-
 template <>
 class [[__sycl_detail__::sycl_type(local_accessor)]] local_accessor <int> {};
 // CHECK: ClassTemplateSpecializationDecl {{.*}} class local_accessor definition

clang/test/SemaSYCL/spec-const-kernel-arg.cpp

@@ -17,11 +17,6 @@ class [[__sycl_detail__::sycl_type()]] B {};
 // expected-error@+1{{'sycl_type' attribute argument 'NotValidType' is not supported}}
 class [[__sycl_detail__::sycl_type(NotValidType)]] C {};
 
-// expected-note@+1{{previous attribute is here}}
-class [[__sycl_detail__::sycl_type(spec_constant)]] spec_constant;
-// expected-error@+1{{attribute 'sycl_type' is already applied with different arguments}}
-class [[__sycl_detail__::sycl_type(accessor)]] spec_constant {};
-
 // expected-error@+2{{attribute 'sycl_type' is already applied with different arguments}}
 // expected-note@+1{{previous attribute is here}}
 class [[__sycl_detail__::sycl_type(group)]] [[__sycl_detail__::sycl_type(accessor)]] group {};