[SYCL][SCLA] Check allocated types are trivial (#13105)

Replace check for cv-unqualified object types with a check for
cv-unqualified trivial types to be in line with the
`sycl_ext_oneapi_private_alloca` extension specification:

> `ElementType` must be a cv-unqualified trivial type

---------

Signed-off-by: Victor Perez <[email protected]>

Author: victor-eds

clang/include/clang/Basic/DiagnosticSemaKinds.td

@@ -191,7 +191,7 @@ def err_intel_sycl_alloca_wrong_arg
             "'sycl::kernel_handler &'. Got %0">;
 def err_intel_sycl_alloca_wrong_type
     : Error<"__builtin_intel_sycl_alloca can only return 'sycl::private_ptr' "
-            "to a cv-unqualified object type. Got %0">;
+            "to a cv-unqualified trivial type. Got %0">;
 def err_intel_sycl_alloca_wrong_size
     : Error<"__builtin_intel_sycl_alloca must be passed a specialization "
             "constant of integral value type as a template argument. Got %1 (%0)">;

clang/lib/Sema/SemaChecking.cpp

@@ -7809,22 +7809,22 @@ bool Sema::CheckIntelSYCLAllocaBuiltinFunctionCall(unsigned, CallExpr *Call) {
 
   // Check the return type is `sycl::multi_ptr<ET,
   // sycl::access::address_space::private_space, DecoratedAddress>`:
-  // - `ET`: non-const, non-volatile, non-void, non-function, non-reference type
-  constexpr auto CheckType = [](QualType RT) {
+  // - `ET`: cv-unqualified trivial type
+  constexpr auto CheckType = [](QualType RT, const ASTContext &Ctx) {
     if (!isSyclType(RT, SYCLTypeAttr::multi_ptr))
       return true;
     // Check element type
     const TemplateArgumentList &TAL =
         cast<ClassTemplateSpecializationDecl>(RT->getAsRecordDecl())
             ->getTemplateArgs();
     QualType ET = TAL.get(0).getAsType();
-    if (ET.isConstQualified() || ET.isVolatileQualified() || ET->isVoidType() ||
-        ET->isFunctionType() || ET->isReferenceType())
+    if (ET.isConstQualified() || ET.isVolatileQualified() ||
+        !ET.isTrivialType(Ctx))
       return true;
     constexpr uint64_t PrivateAS = 0;
     return TAL.get(1).getAsIntegral() != PrivateAS;
   };
-  if (CheckType(FD->getReturnType())) {
+  if (CheckType(FD->getReturnType(), getASTContext())) {
     Diag(Loc, diag::err_intel_sycl_alloca_wrong_type) << FD->getReturnType();
     return true;
   }

clang/test/SemaSYCL/builtin-alloca-errors-device.cpp

@@ -12,6 +12,8 @@ constexpr sycl::specialization_id<float> badsize(1);
 
 struct wrapped_int { int a; };
 
+struct non_trivial { int a = 1; };
+
 template <typename ElementType, auto &Size,
           sycl::access::decorated DecorateAddress>
 __SYCL_BUILTIN_ALIAS(__builtin_intel_sycl_alloca)
@@ -88,24 +90,27 @@ void test(sycl::kernel_handler &h) {
   // expected-error@+1 {{__builtin_intel_sycl_alloca expects to be passed an argument of type 'sycl::kernel_handler &'. Got 'const sycl::kernel_handler &'}}
   private_alloca_bad_5<float, size, sycl::access::decorated::yes>(h);
 
-  // expected-error@+1 {{__builtin_intel_sycl_alloca can only return 'sycl::private_ptr' to a cv-unqualified object type. Got 'multi_ptr<const float, access::address_space::private_space, (decorated)0>'}}
+  // expected-error@+1 {{__builtin_intel_sycl_alloca can only return 'sycl::private_ptr' to a cv-unqualified trivial type. Got 'multi_ptr<const float, access::address_space::private_space, (decorated)0>'}}
   sycl::ext::oneapi::experimental::private_alloca<const float, size, sycl::access::decorated::no>(h);
 
-  // expected-error@+1 {{__builtin_intel_sycl_alloca can only return 'sycl::private_ptr' to a cv-unqualified object type. Got 'multi_ptr<volatile float, access::address_space::private_space, (decorated)0>'}}
+  // expected-error@+1 {{__builtin_intel_sycl_alloca can only return 'sycl::private_ptr' to a cv-unqualified trivial type. Got 'multi_ptr<volatile float, access::address_space::private_space, (decorated)0>'}}
   sycl::ext::oneapi::experimental::private_alloca<volatile float, size, sycl::access::decorated::no>(h);
 
-  // expected-error@+1 {{__builtin_intel_sycl_alloca can only return 'sycl::private_ptr' to a cv-unqualified object type. Got 'multi_ptr<void, access::address_space::private_space, (decorated)1>'}}
+  // expected-error@+1 {{__builtin_intel_sycl_alloca can only return 'sycl::private_ptr' to a cv-unqualified trivial type. Got 'multi_ptr<void, access::address_space::private_space, (decorated)1>'}}
   sycl::ext::oneapi::experimental::private_alloca<void, size, sycl::access::decorated::yes>(h);
 
-  // expected-error@+1 {{__builtin_intel_sycl_alloca can only return 'sycl::private_ptr' to a cv-unqualified object type. Got 'multi_ptr<int *(int), access::address_space::private_space, (decorated)0>'}}
+  // expected-error@+1 {{__builtin_intel_sycl_alloca can only return 'sycl::private_ptr' to a cv-unqualified trivial type. Got 'multi_ptr<int *(int), access::address_space::private_space, (decorated)0>'}}
   sycl::ext::oneapi::experimental::private_alloca<int *(int), size, sycl::access::decorated::no>(h);
 
-  // expected-error@+1 {{__builtin_intel_sycl_alloca can only return 'sycl::private_ptr' to a cv-unqualified object type. Got 'multi_ptr<int &, access::address_space::private_space, (decorated)0>'}}
+  // expected-error@+1 {{__builtin_intel_sycl_alloca can only return 'sycl::private_ptr' to a cv-unqualified trivial type. Got 'multi_ptr<int &, access::address_space::private_space, (decorated)0>'}}
   sycl::ext::oneapi::experimental::private_alloca<int &, size, sycl::access::decorated::no>(h);
 
-  // expected-error@+1 {{__builtin_intel_sycl_alloca can only return 'sycl::private_ptr' to a cv-unqualified object type. Got 'sycl::multi_ptr<float, sycl::access::address_space::local_space, (decorated)0>'}}
+  // expected-error@+1 {{__builtin_intel_sycl_alloca can only return 'sycl::private_ptr' to a cv-unqualified trivial type. Got 'sycl::multi_ptr<float, sycl::access::address_space::local_space, (decorated)0>'}}
   private_alloca_bad_6<float, size, sycl::access::decorated::no>(h);
 
+  // expected-error@+1 {{__builtin_intel_sycl_alloca can only return 'sycl::private_ptr' to a cv-unqualified trivial type. Got 'multi_ptr<non_trivial, access::address_space::private_space, (decorated)1>'}}
+  sycl::ext::oneapi::experimental::private_alloca<non_trivial, size, sycl::access::decorated::yes>(h);
+
   // expected-error@+1 {{__builtin_intel_sycl_alloca must be passed a specialization constant of integral value type as a template argument. Got 'int'}}
   private_alloca_bad_7<float, int, sycl::access::decorated::no>(h);