[SYCL] Turning off the deprecated attribute propagation and Sycl2017Compat option specification (#14984)

AndreiZibrov · web-flow · commit b8615ffe6065 · 2024-08-14T17:01:04.000Z
As a continuation of [#14544](#14544) this PR turns off attribute propagation specified by sycl-1.2.1. It also removes rest of 2017/121 things like a definition of Sycl2017Compat which had been turned off in [#14239](#14239)
diff --git a/clang/include/clang/Basic/DiagnosticGroups.td b/clang/include/clang/Basic/DiagnosticGroups.td
@@ -1375,9 +1375,8 @@ def OpenMP : DiagGroup<"openmp", [
   ]>;
 
 // SYCL warnings
-def Sycl2017Compat : DiagGroup<"sycl-2017-compat">;
 def Sycl2020Compat : DiagGroup<"sycl-2020-compat">;
-def SyclStrict : DiagGroup<"sycl-strict", [ Sycl2017Compat, Sycl2020Compat]>;
+def SyclStrict : DiagGroup<"sycl-strict", [ Sycl2020Compat]>;
 def SyclTarget : DiagGroup<"sycl-target">;
 def SyclFPGAMismatch : DiagGroup<"sycl-fpga-mismatch">;
 def SyclAspectMismatch : DiagGroup<"sycl-aspect-mismatch">;
diff --git a/clang/include/clang/Basic/DiagnosticSemaKinds.td b/clang/include/clang/Basic/DiagnosticSemaKinds.td
@@ -12485,14 +12485,6 @@ def err_sycl_invalid_aspect_argument : Error<
 def warn_sycl_pass_by_value_deprecated 
     : Warning<"passing kernel functions by value is deprecated in SYCL 2020">,
       InGroup<Sycl2020Compat>, ShowInSystemHeader;
-def warn_sycl_pass_by_reference_future 
-    : Warning<"passing of kernel functions by reference is a SYCL 2020 extension">,
-      InGroup<Sycl2017Compat>, ShowInSystemHeader;
-def warn_sycl_implicit_decl
-    : Warning<"SYCL 1.2.1 specification requires an explicit forward "
-              "declaration for a kernel type name; your program may not "
-              "be portable">,
-      InGroup<SyclStrict>, ShowInSystemHeader, DefaultIgnore;
 def warn_sycl_potentially_invalid_as_cast : Warning<
   "explicit cast from %0 to %1 potentially leads to an invalid address space"
   " cast in the resulting code">, InGroup<SyclStrict>,
@@ -12528,9 +12520,6 @@ def err_sycl_mismatch_group_size
 def note_sycl_kernel_declared_here : Note<"kernel declared here">;
 def err_sycl_expected_finalize_method : Error<
   "expected a 'finalize' method for the 'stream' class">;
-def ext_sycl_2020_attr_spelling : ExtWarn<
-  "use of attribute %0 is a SYCL 2020 extension">,
-  InGroup<Sycl2017Compat>;
 def err_sycl_esimd_not_supported_for_type : Error<
   "type %0 is not supported in ESIMD context">;
 def err_sycl_taking_address_of_wrong_function : Error<
diff --git a/clang/include/clang/Basic/LangOptions.h b/clang/include/clang/Basic/LangOptions.h
@@ -146,15 +146,13 @@ class LangOptionsBase {
 
   enum SYCLMajorVersion {
     SYCL_None,
-    SYCL_2017,
     SYCL_2020,
     // The "default" SYCL version to be used when none is specified on the
     // frontend command line.
     SYCL_Default = SYCL_2020
   };
 
   enum class SYCLVersionList {
-    sycl_1_2_1,
     undefined
   };
 
diff --git a/clang/lib/Sema/SemaDeclAttr.cpp b/clang/lib/Sema/SemaDeclAttr.cpp
@@ -151,8 +151,8 @@ void Sema::CheckDeprecatedSYCLAttributeSpelling(const ParsedAttr &A,
     return;
   }
 
-  // Diagnose SYCL 2017 spellings in later SYCL modes.
-  if (LangOpts.getSYCLVersion() > LangOptions::SYCL_2017) {
+  // Diagnose SYCL 2020 spellings in later SYCL modes.
+  if (LangOpts.getSYCLVersion() >= LangOptions::SYCL_2020) {
     // All attributes in the cl vendor namespace are deprecated in favor of a
     // name in the sycl namespace as of SYCL 2020.
     if (A.hasScope() && A.getScopeName()->isStr("cl")) {
@@ -177,14 +177,6 @@ void Sema::CheckDeprecatedSYCLAttributeSpelling(const ParsedAttr &A,
       return;
     }
   }
-
-  // Diagnose SYCL 2020 spellings used in earlier SYCL modes as being an
-  // extension.
-  if (LangOpts.getSYCLVersion() == LangOptions::SYCL_2017 && A.hasScope() &&
-      A.getScopeName()->isStr("sycl")) {
-    Diag(A.getLoc(), diag::ext_sycl_2020_attr_spelling) << A;
-    return;
-  }
 }
 
 /// Check if IdxExpr is a valid parameter index for a function or
@@ -4358,19 +4350,13 @@ static void handleSYCLIntelLoopFuseAttr(Sema &S, Decl *D, const ParsedAttr &A) {
 }
 
 static void handleVecTypeHint(Sema &S, Decl *D, const ParsedAttr &AL) {
-  // This attribute is deprecated without replacement in SYCL 2020 mode.
+  // Given attribute is deprecated without replacement in SYCL 2020 mode.
   // Ignore the attribute in SYCL 2020.
-  if (S.LangOpts.getSYCLVersion() > LangOptions::SYCL_2017) {
+  if (S.LangOpts.getSYCLVersion() >= LangOptions::SYCL_2020) {
     S.Diag(AL.getLoc(), diag::warn_attribute_deprecated_ignored) << AL;
     return;
   }
 
-  // If the attribute is used with the [[sycl::vec_type_hint]] spelling in SYCL
-  // 2017 mode, we want to warn about using the newer name in the older
-  // standard as a compatibility extension.
-  if (S.LangOpts.getSYCLVersion() == LangOptions::SYCL_2017 && AL.hasScope())
-    S.Diag(AL.getLoc(), diag::ext_sycl_2020_attr_spelling) << AL;
-
   if (!AL.hasParsedType()) {
     S.Diag(AL.getLoc(), diag::err_attribute_wrong_number_arguments) << AL << 1;
     return;
diff --git a/clang/lib/Sema/SemaSYCL.cpp b/clang/lib/Sema/SemaSYCL.cpp
@@ -539,11 +539,9 @@ static void collectSYCLAttributes(SemaSYCL &S, FunctionDecl *FD,
   if (!FD->hasAttrs())
     return;
 
-  // In SYCL 1.2.1 mode, the attributes are propagated from the function they
-  // are applied to onto the kernel which calls the function.
-  // In SYCL 2020 mode, the attributes are not propagated to the kernel.
-  if (DirectlyCalled || S.getASTContext().getLangOpts().getSYCLVersion() <
-                            LangOptions::SYCL_2020) {
+  // In SYCL 2020 mode, the attributes aren't propagated from the function they
+  // are applied on to the kernel which calls the function.
+  if (DirectlyCalled) {
     llvm::copy_if(FD->getAttrs(), std::back_inserter(Attrs), [](Attr *A) {
       // FIXME: Make this list self-adapt as new SYCL attributes are added.
       return isa<IntelReqdSubGroupSizeAttr, IntelNamedSubGroupSizeAttr,
@@ -553,14 +551,8 @@ static void collectSYCLAttributes(SemaSYCL &S, FunctionDecl *FD,
                  SYCLIntelMaxWorkGroupSizeAttr, SYCLIntelMaxGlobalWorkDimAttr,
                  SYCLIntelMinWorkGroupsPerComputeUnitAttr,
                  SYCLIntelMaxWorkGroupsPerMultiprocessorAttr,
-                 SYCLIntelNoGlobalWorkOffsetAttr, SYCLSimdAttr>(A);
-    });
-  }
-
-  // Attributes that should not be propagated from device functions to a kernel.
-  if (DirectlyCalled) {
-    llvm::copy_if(FD->getAttrs(), std::back_inserter(Attrs), [](Attr *A) {
-      return isa<SYCLIntelLoopFuseAttr, SYCLIntelMaxConcurrencyAttr,
+                 SYCLIntelNoGlobalWorkOffsetAttr, SYCLSimdAttr,
+                 SYCLIntelLoopFuseAttr, SYCLIntelMaxConcurrencyAttr,
                  SYCLIntelDisableLoopPipeliningAttr,
                  SYCLIntelInitiationIntervalAttr,
                  SYCLIntelUseStallEnableClustersAttr, SYCLDeviceHasAttr,
@@ -4856,10 +4848,6 @@ class SYCLKernelNameTypeVisitor
                       scope */
                 0 << KernelNameType;
             IsInvalid = true;
-          } else {
-            S.Diag(KernelInvocationFuncLoc, diag::warn_sycl_implicit_decl);
-            S.Diag(DeclNamed->getLocation(), diag::note_previous_decl)
-                << DeclNamed->getName();
           }
         }
       }
@@ -4930,13 +4918,9 @@ void SemaSYCL::CheckSYCLKernelCall(FunctionDecl *KernelFunc,
 
   // check that calling kernel conforms to spec
   QualType KernelParamTy = KernelFunc->getParamDecl(0)->getType();
-  if (KernelParamTy->isReferenceType()) {
-    // passing by reference, so emit warning if not using SYCL 2020
-    if (SemaRef.LangOpts.getSYCLVersion() < LangOptions::SYCL_2020)
-      Diag(KernelFunc->getLocation(), diag::warn_sycl_pass_by_reference_future);
-  } else {
+  if (not KernelParamTy->isReferenceType()) {
     // passing by value.  emit warning if using SYCL 2020 or greater
-    if (SemaRef.LangOpts.getSYCLVersion() > LangOptions::SYCL_2017)
+    if (SemaRef.LangOpts.getSYCLVersion() >= LangOptions::SYCL_2020)
       Diag(KernelFunc->getLocation(), diag::warn_sycl_pass_by_value_deprecated);
   }
 
diff --git a/clang/test/SemaSYCL/implicit_kernel_type.cpp b/clang/test/SemaSYCL/implicit_kernel_type.cpp
@@ -29,32 +29,6 @@ int main() {
     h.single_task<InvalidKernelName1>([]() {});
   });
 
-#if defined(WARN)
-  // expected-warning@#KernelSingleTask {{SYCL 1.2.1 specification requires an explicit forward declaration for a kernel type name; your program may not be portable}}
-#elif defined(ERROR)
-  // expected-error@#KernelSingleTask {{SYCL 1.2.1 specification requires an explicit forward declaration for a kernel type name; your program may not be portable}}
-#endif
-
-  q.submit([&](handler &h) {
-  // expected-note@+2 {{fake_kernel declared here}}
-  // expected-note@+1 {{in instantiation of function template specialization}}
-    h.single_task<class fake_kernel>([]() { function(); });
-  });
-
-#if defined(WARN)
-  // expected-warning@#KernelSingleTask {{SYCL 1.2.1 specification requires an explicit forward declaration for a kernel type name; your program may not be portable}}
-#elif defined(ERROR)
-  // expected-error@#KernelSingleTask {{SYCL 1.2.1 specification requires an explicit forward declaration for a kernel type name; your program may not be portable}}
-#endif
-
-  q.submit([&](handler &h) {
-  // expected-note@+2 {{fake_kernel2 declared here}}
-  // expected-note@+1 {{in instantiation of function template specialization}}
-    h.single_task<class fake_kernel2>([]() {
-      auto l = [](auto f) { f(); };
-    });
-  });
-
   q.submit([&](handler &h) {
     h.single_task<class myWrapper>([]() { function(); });
   });
