From 187616295aa358e322ec9444a0c4549a8e999da9 Mon Sep 17 00:00:00 2001
From: Andrei Elovikov <andrei.elovikov@intel.com>
Date: Thu, 12 Dec 2024 14:02:21 -0800
Subject: [PATCH 1/2] [SYCL][E2E] Add a test showing pre-existing
 swizzle-related bugs

---
 sycl/test/basic_tests/vectors/swizzle.cpp | 43 +++++++++++++++++++++++
 1 file changed, 43 insertions(+)
 create mode 100644 sycl/test/basic_tests/vectors/swizzle.cpp

diff --git a/sycl/test/basic_tests/vectors/swizzle.cpp b/sycl/test/basic_tests/vectors/swizzle.cpp
new file mode 100644
index 0000000000000..70da87e8f1964
--- /dev/null
+++ b/sycl/test/basic_tests/vectors/swizzle.cpp
@@ -0,0 +1,43 @@
+// RUN: %clangxx -fsycl %s -o %t_default.out
+// RUN: %t_default.out
+
+// FIXME: Everything should compile cleanly.
+// RUN: %clangxx -fsycl -fsycl-device-only -DCHECK_ERRORS -fsyntax-only -Xclang -verify -Xclang -verify-ignore-unexpected=note,error %s
+
+#include <sycl/vector.hpp>
+
+int main() {
+  sycl::vec<int, 4> v{1, 2, 3, 4};
+  auto sw = v.swizzle<1, 2>();
+  assert(sw.lo()[0] == 2);
+  assert(sw.hi()[0] == 3);
+
+  // FIXME: Should be "4":
+  assert((sw + sw).lo()[0] == 2);
+
+  // FIXME: The below should compile.
+#if CHECK_ERRORS
+  // expected-error-re@+1 {{no template named 'swizzle' in {{.*}}}}
+  assert(sw.swizzle<0>()[0] == 2);
+  // expected-error-re@+1 {{no template named 'swizzle' in {{.*}}}}
+  assert(sw.swizzle<1>()[0] == 3);
+
+  {
+    // expected-error-re@+1 {{no template named 'swizzle' in {{.*}}}}
+    auto tmp = sw.swizzle<1, 0>();
+    assert(tmp[0] == 3);
+    assert(tmp[1] == 2);
+  }
+
+  {
+    // expected-error-re@+1 {{no template named 'swizzle' in {{.*}}}}
+    auto tmp = (sw + sw).swizzle<1, 0>();
+    std::cout << tmp[0] << " " << tmp[1] << std::endl;
+
+    assert(tmp[0] == 6);
+    assert(tmp[1] == 4);
+  }
+#endif
+
+  return 0;
+}

From a8639492661783f4ac2b944fb1bf2decab568a51 Mon Sep 17 00:00:00 2001
From: Andrei Elovikov <andrei.elovikov@intel.com>
Date: Thu, 12 Dec 2024 14:15:04 -0800
Subject: [PATCH 2/2] [SYCL] Implement "swizzle" member function on swizzles

It brings it closer to SYCL spec and makes more consistent by having "named"
swizzle member functions (like `.hi()`/`.lo()`) and `.swizzle<...>()` behave in
a similar way, i.e. we still have a bug when doing a swizzle on an expression
tree.

Note that the whole "expression tree" machinery is not standard-conformant and
will be completely removed separately (under `-fpreview-breaking-changes` flag).

I still want to implement this partial bugfix so that I could switch to a
unified mixin-based implementation for swizzles on `vec`/`swizzle` classes,
instead of doing preprocessor tricks with `swizzles.def` file.
---
 sycl/include/sycl/vector.hpp              | 10 ++++++++++
 sycl/test/basic_tests/vectors/swizzle.cpp | 15 +++------------
 2 files changed, 13 insertions(+), 12 deletions(-)

diff --git a/sycl/include/sycl/vector.hpp b/sycl/include/sycl/vector.hpp
index 97d9704c3cc26..75eb80cac829c 100644
--- a/sycl/include/sycl/vector.hpp
+++ b/sycl/include/sycl/vector.hpp
@@ -1207,6 +1207,16 @@ class SwizzleOp {
   };
 
 public:
+  template <int... swizzleIndexes>
+  ConstSwizzle<Indexer<swizzleIndexes>::value...> swizzle() const {
+    return m_Vector;
+  }
+
+  template <int... swizzleIndexes>
+  Swizzle<Indexer<swizzleIndexes>::value...> swizzle() {
+    return m_Vector;
+  }
+
 #ifdef __SYCL_ACCESS_RETURN
 #error "Undefine __SYCL_ACCESS_RETURN macro"
 #endif
diff --git a/sycl/test/basic_tests/vectors/swizzle.cpp b/sycl/test/basic_tests/vectors/swizzle.cpp
index 70da87e8f1964..3b9c217bc1878 100644
--- a/sycl/test/basic_tests/vectors/swizzle.cpp
+++ b/sycl/test/basic_tests/vectors/swizzle.cpp
@@ -1,9 +1,6 @@
 // RUN: %clangxx -fsycl %s -o %t_default.out
 // RUN: %t_default.out
 
-// FIXME: Everything should compile cleanly.
-// RUN: %clangxx -fsycl -fsycl-device-only -DCHECK_ERRORS -fsyntax-only -Xclang -verify -Xclang -verify-ignore-unexpected=note,error %s
-
 #include <sycl/vector.hpp>
 
 int main() {
@@ -15,29 +12,23 @@ int main() {
   // FIXME: Should be "4":
   assert((sw + sw).lo()[0] == 2);
 
-  // FIXME: The below should compile.
-#if CHECK_ERRORS
-  // expected-error-re@+1 {{no template named 'swizzle' in {{.*}}}}
   assert(sw.swizzle<0>()[0] == 2);
-  // expected-error-re@+1 {{no template named 'swizzle' in {{.*}}}}
   assert(sw.swizzle<1>()[0] == 3);
 
   {
-    // expected-error-re@+1 {{no template named 'swizzle' in {{.*}}}}
     auto tmp = sw.swizzle<1, 0>();
     assert(tmp[0] == 3);
     assert(tmp[1] == 2);
   }
 
   {
-    // expected-error-re@+1 {{no template named 'swizzle' in {{.*}}}}
     auto tmp = (sw + sw).swizzle<1, 0>();
     std::cout << tmp[0] << " " << tmp[1] << std::endl;
 
-    assert(tmp[0] == 6);
-    assert(tmp[1] == 4);
+    // FIXME: Should be "6" and "4", respectively.
+    assert(tmp[0] == 3);
+    assert(tmp[1] == 2);
   }
-#endif
 
   return 0;
 }