[SYCL][SCLA][NFC] Fix warning and potential test errors (#13937)

- `DecorateAddressIndex` is not used. Drop to avoid warnings.
- Modify `test` function implementing tests so it does not read from
stdin, but receive its parameter as an argument and read these arguments
from stdin in `main`.
- Manually inline `allocate` function as this usage of
`[aligned_]private_alloca` is undefined behavior

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

Author: victor-eds

clang/lib/CodeGen/CGBuiltin.cpp

@@ -23922,7 +23922,6 @@ RValue CodeGenFunction::EmitIntelSYCLAllocaBuiltin(
   constexpr unsigned ElementTypeIndex = 0;
   const unsigned AlignmentIndex = IsAlignedAlloca ? 1 : InvalidIndex;
   const unsigned SpecNameIndex = IsAlignedAlloca ? 2 : 1;
-  const unsigned DecorateAddressIndex = IsAlignedAlloca ? 3 : 2;
 
   const FunctionDecl *FD = E->getDirectCallee();
   assert(FD && "Expecting direct call to builtin");

sycl/test-e2e/PrivateAlloca/ValidUsage/Inputs/private_alloca_test.hpp

@@ -11,39 +11,29 @@ template <typename ElementType, typename SizeType,
           sycl::access::decorated DecorateAddress, std::size_t Alignment>
 class Kernel;
 
-template <typename ElementType, auto &Size,
-          sycl::access::decorated DecorateAddress, std::size_t Alignment>
-static auto allocate(sycl::kernel_handler &kh) {
-  if constexpr (Alignment > 0) {
-    return sycl::ext::oneapi::experimental::aligned_private_alloca<
-        ElementType, Alignment, Size, DecorateAddress>(kh);
-  } else {
-    return sycl::ext::oneapi::experimental::private_alloca<ElementType, Size,
-                                                           DecorateAddress>(kh);
-  }
-}
-
 template <typename ElementType, auto &Size,
           sycl::access::decorated DecorateAddress, std::size_t Alignment = 0>
-void test() {
-  std::size_t N;
-
-  std::cin >> N;
-
-  std::vector<std::size_t> v(N);
+void test(std::size_t n) {
+  std::vector<std::size_t> v(n);
   {
     sycl::queue q;
     sycl::buffer<std::size_t> b(v);
     q.submit([&](sycl::handler &cgh) {
       sycl::accessor acc(b, cgh, sycl::write_only, sycl::no_init);
-      cgh.set_specialization_constant<Size>(N);
+      cgh.set_specialization_constant<Size>(n);
       using spec_const_type = std::remove_reference_t<decltype(Size)>;
       using size_type = typename spec_const_type::value_type;
       cgh.single_task<
           Kernel<ElementType, size_type, DecorateAddress, Alignment>>(
           [=](sycl::kernel_handler h) {
-            auto ptr =
-                allocate<ElementType, Size, DecorateAddress, Alignment>(h);
+            sycl::private_ptr<ElementType, DecorateAddress> ptr;
+            if constexpr (Alignment > 0) {
+              ptr = sycl::ext::oneapi::experimental::aligned_private_alloca<
+                  ElementType, Alignment, Size, DecorateAddress>(h);
+            } else {
+              ptr = sycl::ext::oneapi::experimental::private_alloca<
+                  ElementType, Size, DecorateAddress>(h);
+            }
             const std::size_t M = h.get_specialization_constant<Size>();
             ptr[0] = static_cast<ElementType>(M);
             ElementType value{1};
@@ -60,9 +50,9 @@ void test() {
     });
     q.wait_and_throw();
   }
-  assert(static_cast<std::size_t>(v.front()) == N &&
+  assert(static_cast<std::size_t>(v.front()) == n &&
          "Wrong private alloca length reported");
-  for (std::size_t i = 1; i < N; ++i) {
+  for (std::size_t i = 1; i < n; ++i) {
     assert(static_cast<std::size_t>(v[i]) == i &&
            "Wrong value in copied-back sequence");
   }

sycl/test-e2e/PrivateAlloca/ValidUsage/private_alloca_bool_size.cpp

@@ -9,6 +9,8 @@
 constexpr sycl::specialization_id<bool> size(true);
 
 int main() {
-  test<int, size, sycl::access::decorated::legacy>();
-  test<int, size, sycl::access::decorated::legacy, alignof(int) * 2>();
+  std::size_t n = 0;
+  std::cin >> n;
+  test<int, size, sycl::access::decorated::legacy>(n);
+  test<int, size, sycl::access::decorated::legacy, alignof(int) * 2>(n);
 }

sycl/test-e2e/PrivateAlloca/ValidUsage/private_alloca_decorated.cpp

@@ -12,6 +12,8 @@
 constexpr sycl::specialization_id<int> size(10);
 
 int main() {
-  test<float, size, sycl::access::decorated::yes>();
-  test<float, size, sycl::access::decorated::yes, alignof(float) * 2>();
+  std::size_t n = 0;
+  std::cin >> n;
+  test<float, size, sycl::access::decorated::yes>(n);
+  test<float, size, sycl::access::decorated::yes, alignof(float) * 2>(n);
 }

sycl/test-e2e/PrivateAlloca/ValidUsage/private_alloca_legacy.cpp

@@ -12,6 +12,8 @@
 constexpr sycl::specialization_id<int16_t> size(10);
 
 int main() {
-  test<int, size, sycl::access::decorated::legacy>();
-  test<int, size, sycl::access::decorated::legacy, alignof(int) * 4>();
+  std::size_t n = 0;
+  std::cin >> n;
+  test<int, size, sycl::access::decorated::legacy>(n);
+  test<int, size, sycl::access::decorated::legacy, alignof(int) * 4>(n);
 }

sycl/test-e2e/PrivateAlloca/ValidUsage/private_alloca_multiple.cpp

@@ -11,11 +11,20 @@ constexpr sycl::specialization_id<int> isize(10);
 constexpr sycl::specialization_id<int16_t> ssize(100);
 
 int main() {
-  test<float, size, sycl::access::decorated::yes>();
-  test<int16_t, isize, sycl::access::decorated::legacy>();
-  test<int, ssize, sycl::access::decorated::no>();
+  constexpr std::size_t num_tests = 3;
+  std::array<std::size_t, num_tests> ns;
+  std::generate_n(ns.begin(), num_tests, []() {
+    std::size_t i = 0;
+    std::cin >> i;
+    return i;
+  });
 
-  test<float, size, sycl::access::decorated::yes, alignof(float) * 2>();
-  test<int16_t, isize, sycl::access::decorated::legacy, alignof(int16_t) * 2>();
-  test<int, ssize, sycl::access::decorated::no, alignof(int) * 2>();
+  test<float, size, sycl::access::decorated::yes>(ns[0]);
+  test<int16_t, isize, sycl::access::decorated::legacy>(ns[1]);
+  test<int, ssize, sycl::access::decorated::no>(ns[2]);
+
+  test<float, size, sycl::access::decorated::yes, alignof(float) * 2>(ns[0]);
+  test<int16_t, isize, sycl::access::decorated::legacy, alignof(int16_t) * 2>(
+      ns[1]);
+  test<int, ssize, sycl::access::decorated::no, alignof(int) * 2>(ns[2]);
 }

sycl/test-e2e/PrivateAlloca/ValidUsage/private_alloca_raw.cpp

@@ -57,7 +57,10 @@ class value_and_sign {
 };
 
 int main() {
-  test<value_and_sign, size, sycl::access::decorated::no>();
+  std::size_t n = 0;
+  std::cin >> n;
+
+  test<value_and_sign, size, sycl::access::decorated::no>(n);
   test<value_and_sign, size, sycl::access::decorated::no,
-       alignof(value_and_sign) * 2>();
+       alignof(value_and_sign) * 2>(n);
 }