From 950c0838d3a672a267ec06ecf42b3dd7d37a2057 Mon Sep 17 00:00:00 2001
From: "Larsen, Steffen" <steffen.larsen@intel.com>
Date: Wed, 13 Mar 2024 08:37:18 -0700
Subject: [PATCH 1/6] [SYCL] Make swizzle mutating operators const friends

In https://github.com/intel/llvm/pull/12682 the mutating operators for
swizzles (+=, -=, ..., ++, --) were reverted to be members rather than
friends. Since swizzles mutate the underlying vec rather than themselves
these operators should take and return constant references instead,
which this commit implements.

Signed-off-by: Larsen, Steffen <steffen.larsen@intel.com>
---
 sycl/include/sycl/types.hpp                   | 35 ++++++++-
 sycl/test-e2e/Regression/swizzle_opassign.cpp | 78 +++++++++++++++++++
 2 files changed, 112 insertions(+), 1 deletion(-)
 create mode 100644 sycl/test-e2e/Regression/swizzle_opassign.cpp

diff --git a/sycl/include/sycl/types.hpp b/sycl/include/sycl/types.hpp
index 7d904a8246b03..4562f8fad4d83 100644
--- a/sycl/include/sycl/types.hpp
+++ b/sycl/include/sycl/types.hpp
@@ -1849,6 +1849,25 @@ class SwizzleOp {
 #ifdef __SYCL_OPASSIGN
 #error "Undefine __SYCL_OPASSIGN macro."
 #endif
+#ifdef __INTEL_PREVIEW_BREAKING_CHANGES
+#define __SYCL_OPASSIGN(OPASSIGN, OP)                                          \
+  friend const SwizzleOp &operator OPASSIGN(const SwizzleOp & Lhs,             \
+                                            const DataT & Rhs) {               \
+    Lhs.operatorHelper<OP>(vec_t(Rhs));                                        \
+    return Lhs;                                                                \
+  }                                                                            \
+  template <typename RhsOperation>                                             \
+  friend const SwizzleOp &operator OPASSIGN(const SwizzleOp & Lhs,             \
+                                            const RhsOperation & Rhs) {        \
+    Lhs.operatorHelper<OP>(Rhs);                                               \
+    return Lhs;                                                                \
+  }                                                                            \
+  friend const SwizzleOp &operator OPASSIGN(const SwizzleOp & Lhs,             \
+                                            const vec_t & Rhs) {               \
+    Lhs.operatorHelper<OP>(Rhs);                                               \
+    return Lhs;                                                                \
+  }
+#else // __INTEL_PREVIEW_BREAKING_CHANGES
 #define __SYCL_OPASSIGN(OPASSIGN, OP)                                          \
   SwizzleOp &operator OPASSIGN(const DataT & Rhs) {                            \
     operatorHelper<OP>(vec_t(Rhs));                                            \
@@ -1859,6 +1878,7 @@ class SwizzleOp {
     operatorHelper<OP>(Rhs);                                                   \
     return *this;                                                              \
   }
+#endif // __INTEL_PREVIEW_BREAKING_CHANGES
 
   __SYCL_OPASSIGN(+=, std::plus)
   __SYCL_OPASSIGN(-=, std::minus)
@@ -1875,6 +1895,18 @@ class SwizzleOp {
 #ifdef __SYCL_UOP
 #error "Undefine __SYCL_UOP macro"
 #endif
+#ifdef __INTEL_PREVIEW_BREAKING_CHANGES
+#define __SYCL_UOP(UOP, OPASSIGN)                                              \
+  friend const SwizzleOp &operator UOP(const SwizzleOp & sv) {                 \
+    sv OPASSIGN static_cast<DataT>(1);                                         \
+    return sv;                                                                 \
+  }                                                                            \
+  friend vec_t operator UOP(const SwizzleOp &sv, int) {                        \
+    vec_t Ret = sv;                                                            \
+    sv OPASSIGN static_cast<DataT>(1);                                         \
+    return Ret;                                                                \
+  }
+#else // __INTEL_PREVIEW_BREAKING_CHANGES
 #define __SYCL_UOP(UOP, OPASSIGN)                                              \
   SwizzleOp &operator UOP() {                                                  \
     *this OPASSIGN static_cast<DataT>(1);                                      \
@@ -1885,6 +1917,7 @@ class SwizzleOp {
     *this OPASSIGN static_cast<DataT>(1);                                      \
     return Ret;                                                                \
   }
+#endif // __INTEL_PREVIEW_BREAKING_CHANGES
 
   __SYCL_UOP(++, +=)
   __SYCL_UOP(--, -=)
@@ -2416,7 +2449,7 @@ class SwizzleOp {
   }
 
   template <template <typename> class Operation, typename RhsOperation>
-  void operatorHelper(const RhsOperation &Rhs) {
+  void operatorHelper(const RhsOperation &Rhs) const {
     Operation<vec_data_t<DataT>> Op;
     std::array<int, getNumElements()> Idxs{Indexes...};
     for (size_t I = 0; I < Idxs.size(); ++I) {
diff --git a/sycl/test-e2e/Regression/swizzle_opassign.cpp b/sycl/test-e2e/Regression/swizzle_opassign.cpp
new file mode 100644
index 0000000000000..3a6a2357ffdc4
--- /dev/null
+++ b/sycl/test-e2e/Regression/swizzle_opassign.cpp
@@ -0,0 +1,78 @@
+// REQUIRES: aspect-usm_shared_allocations
+// RUN: %{build} -o %t.out
+// RUN: %{run} %t.out
+// RUN: %if preview-breaking-changes-supported %{  %{build} -fpreview-breaking-changes -o %t2.out   %}
+// RUN: %if preview-breaking-changes-supported %{ %{run} %t2.out  %}
+
+// Tests that the mutating operators (+=, -=, ..., ++, --) on swizzles compile
+// and correctly mutate the elements in the corresponding vector.
+
+#include <sycl/sycl.hpp>
+
+constexpr size_t NumOps = 14;
+constexpr std::string_view OpNames[NumOps] = {
+    "+=", "-=",  "*=",  "/=",        "%=",        "&=",         "|=",
+    "^=", "<<=", ">>=", "prefix ++", "prefix --", "postfix ++", "prefix ++"};
+
+int main() {
+  sycl::queue Q;
+  bool *Results = sycl::malloc_shared<bool>(NumOps, Q);
+  for (size_t I = 0; I < NumOps; ++I)
+    Results[I] = 0;
+
+  Q.single_task([=]() {
+     bool *ResultIt = Results;
+#define TestCase(OP)                                                           \
+  {                                                                            \
+    sycl::vec<int, 4> VecVal{1, 2, 3, 4};                                      \
+    int ExpectedRes = VecVal[1] OP 2;                                          \
+    *(ResultIt++) = (VecVal.swizzle<1>() OP## = 2)[0] == ExpectedRes &&        \
+                    VecVal[1] == ExpectedRes;                                  \
+  }
+     TestCase(+);
+     TestCase(-);
+     TestCase(*);
+     TestCase(/);
+     TestCase(%);
+     TestCase(&);
+     TestCase(|);
+     TestCase(^);
+     TestCase(<<);
+     TestCase(>>);
+     {
+       sycl::vec<int, 4> VecVal{1, 2, 3, 4};
+       int ExpectedRes = VecVal[1] + 1;
+       *(ResultIt++) = (++VecVal.swizzle<1>())[0] == ExpectedRes &&
+                       VecVal[1] == ExpectedRes;
+     }
+     {
+       sycl::vec<int, 4> VecVal{1, 2, 3, 4};
+       int ExpectedRes = VecVal[1] - 1;
+       *(ResultIt++) = (--VecVal.swizzle<1>())[0] == ExpectedRes &&
+                       VecVal[1] == ExpectedRes;
+     }
+     {
+       sycl::vec<int, 4> VecVal{1, 2, 3, 4};
+       int ExpectedRes = VecVal[1] + 1;
+       *(ResultIt++) = (VecVal.swizzle<1>()++)[0] == (ExpectedRes - 1) &&
+                       VecVal[1] == ExpectedRes;
+     }
+     {
+       sycl::vec<int, 4> VecVal{1, 2, 3, 4};
+       int ExpectedRes = VecVal[1] - 1;
+       *(ResultIt++) = (VecVal.swizzle<1>()--)[0] == (ExpectedRes + 1) &&
+                       VecVal[1] == ExpectedRes;
+     }
+   }).wait_and_throw();
+
+  int Failures = 0;
+  for (size_t I = 0; I < NumOps; ++I) {
+    if (!Results[I]) {
+      std::cout << "Failed for " << OpNames[I] << std::endl;
+      ++Failures;
+    }
+  }
+
+  sycl::free(Results, Q);
+  return Failures;
+}

From 3262f89009469cd0670ca8689f3c46530b06d726 Mon Sep 17 00:00:00 2001
From: "Larsen, Steffen" <steffen.larsen@intel.com>
Date: Thu, 14 Mar 2024 01:36:52 -0700
Subject: [PATCH 2/6] [SYCL] Make mutable swizzle functions unusable on const
 vec

Swizzles should not expose mutating functions when the underlying vector
is const. This commit SFINAEs these out.

This is built on top of https://github.com/intel/llvm/pull/13012.

Signed-off-by: Larsen, Steffen <steffen.larsen@intel.com>
---
 sycl/include/sycl/types.hpp                   | 93 +++++++++++--------
 .../vectors/const_swizzle_negative.cpp        | 91 ++++++++++++++++++
 2 files changed, 145 insertions(+), 39 deletions(-)
 create mode 100644 sycl/test/basic_tests/vectors/const_swizzle_negative.cpp

diff --git a/sycl/include/sycl/types.hpp b/sycl/include/sycl/types.hpp
index 4562f8fad4d83..e3be3ed524d84 100644
--- a/sycl/include/sycl/types.hpp
+++ b/sycl/include/sycl/types.hpp
@@ -1735,6 +1735,10 @@ class SwizzleOp {
       DataT, std::common_type_t<OpLeftDataT, OpRightDataT>>;
   static constexpr int getNumElements() { return sizeof...(Indexes); }
 
+  template <typename RelayVecT = VecT>
+  static constexpr bool VecIsMutable =
+      !std::is_const_v<RelayVecT> && std::is_same_v<RelayVecT, VecT>;
+
   using rel_t = detail::rel_t<DataT>;
   using vec_t = vec<DataT, sizeof...(Indexes)>;
   using vec_rel_t = vec<rel_t, sizeof...(Indexes)>;
@@ -1760,13 +1764,20 @@ class SwizzleOp {
                                       OperationCurrentT, Indexes...>,
                             OperationCurrentT_, Idx_...>;
 
+  template <int IdxNum>
+  static constexpr bool HasOneIndex =
+      1 == IdxNum && SwizzleOp::getNumElements() == IdxNum;
+
   template <int IdxNum, typename T = void>
-  using EnableIfOneIndex = typename std::enable_if_t<
-      1 == IdxNum && SwizzleOp::getNumElements() == IdxNum, T>;
+  using EnableIfOneIndex = typename std::enable_if_t<HasOneIndex<IdxNum>, T>;
+
+  template <int IdxNum>
+  static constexpr bool HasMultipleIndices =
+      1 != IdxNum && SwizzleOp::getNumElements() == IdxNum;
 
   template <int IdxNum, typename T = void>
-  using EnableIfMultipleIndexes = typename std::enable_if_t<
-      1 != IdxNum && SwizzleOp::getNumElements() == IdxNum, T>;
+  using EnableIfMultipleIndexes =
+      typename std::enable_if_t<HasMultipleIndices<IdxNum>, T>;
 
   template <typename T>
   using EnableIfScalarType = typename std::enable_if_t<
@@ -1851,19 +1862,21 @@ class SwizzleOp {
 #endif
 #ifdef __INTEL_PREVIEW_BREAKING_CHANGES
 #define __SYCL_OPASSIGN(OPASSIGN, OP)                                          \
-  friend const SwizzleOp &operator OPASSIGN(const SwizzleOp & Lhs,             \
-                                            const DataT & Rhs) {               \
+  template <typename RelayVecT = VecT>                                         \
+  friend const std::enable_if_t<VecIsMutable<RelayVecT>, SwizzleOp>            \
+      &operator OPASSIGN(const SwizzleOp & Lhs, const DataT & Rhs) {           \
     Lhs.operatorHelper<OP>(vec_t(Rhs));                                        \
     return Lhs;                                                                \
   }                                                                            \
-  template <typename RhsOperation>                                             \
-  friend const SwizzleOp &operator OPASSIGN(const SwizzleOp & Lhs,             \
-                                            const RhsOperation & Rhs) {        \
+  template <typename RhsOperation, typename RelayVecT = VecT>                  \
+  friend const std::enable_if_t<VecIsMutable<RelayVecT>, SwizzleOp>            \
+      &operator OPASSIGN(const SwizzleOp & Lhs, const RhsOperation & Rhs) {    \
     Lhs.operatorHelper<OP>(Rhs);                                               \
     return Lhs;                                                                \
   }                                                                            \
-  friend const SwizzleOp &operator OPASSIGN(const SwizzleOp & Lhs,             \
-                                            const vec_t & Rhs) {               \
+  template <typename RelayVecT = VecT>                                         \
+  friend const std::enable_if_t<VecIsMutable<RelayVecT>, SwizzleOp>            \
+      &operator OPASSIGN(const SwizzleOp & Lhs, const vec_t & Rhs) {           \
     Lhs.operatorHelper<OP>(Rhs);                                               \
     return Lhs;                                                                \
   }
@@ -1897,11 +1910,15 @@ class SwizzleOp {
 #endif
 #ifdef __INTEL_PREVIEW_BREAKING_CHANGES
 #define __SYCL_UOP(UOP, OPASSIGN)                                              \
-  friend const SwizzleOp &operator UOP(const SwizzleOp & sv) {                 \
+  template <typename RelayVecT = VecT>                                         \
+  friend const std::enable_if_t<VecIsMutable<RelayVecT>, SwizzleOp>            \
+      &operator UOP(const SwizzleOp & sv) {                                    \
     sv OPASSIGN static_cast<DataT>(1);                                         \
     return sv;                                                                 \
   }                                                                            \
-  friend vec_t operator UOP(const SwizzleOp &sv, int) {                        \
+  template <typename RelayVecT = VecT>                                         \
+  friend std::enable_if_t<VecIsMutable<RelayVecT>, vec_t> operator UOP(        \
+      const SwizzleOp & sv, int) {                                             \
     vec_t Ret = sv;                                                            \
     sv OPASSIGN static_cast<DataT>(1);                                         \
     return Ret;                                                                \
@@ -2043,9 +2060,10 @@ class SwizzleOp {
 #undef __SYCL_RELLOGOP
 #endif // defined(__INTEL_PREVIEW_BREAKING_CHANGES)
 
-  template <int IdxNum = getNumElements(),
-            typename = EnableIfMultipleIndexes<IdxNum>>
-  SwizzleOp &operator=(const vec<DataT, IdxNum> &Rhs) {
+  template <int IdxNum = getNumElements(), typename RelayVecT = VecT>
+  std::enable_if_t<HasMultipleIndices<IdxNum> && VecIsMutable<RelayVecT>,
+                   SwizzleOp> &
+  operator=(const vec<DataT, IdxNum> &Rhs) {
     std::array<int, IdxNum> Idxs{Indexes...};
     for (size_t I = 0; I < Idxs.size(); ++I) {
       m_Vector->setValue(Idxs[I], Rhs.getValue(I));
@@ -2053,25 +2071,22 @@ class SwizzleOp {
     return *this;
   }
 
-  template <int IdxNum = getNumElements(), typename = EnableIfOneIndex<IdxNum>>
-  SwizzleOp &operator=(const DataT &Rhs) {
-    std::array<int, IdxNum> Idxs{Indexes...};
-    m_Vector->setValue(Idxs[0], Rhs);
-    return *this;
-  }
-
-  template <int IdxNum = getNumElements(),
-            EnableIfMultipleIndexes<IdxNum, bool> = true>
-  SwizzleOp &operator=(const DataT &Rhs) {
-    std::array<int, IdxNum> Idxs{Indexes...};
-    for (auto Idx : Idxs) {
-      m_Vector->setValue(Idx, Rhs);
+  template <typename RelayVecT = VecT>
+  std::enable_if_t<VecIsMutable<RelayVecT>, SwizzleOp> &
+  operator=(const DataT &Rhs) {
+    std::array<int, getNumElements()> Idxs{Indexes...};
+    if constexpr (getNumElements() == 1) {
+      m_Vector->setValue(Idxs[0], Rhs);
+    } else {
+      for (auto Idx : Idxs)
+        m_Vector->setValue(Idx, Rhs);
     }
     return *this;
   }
 
-  template <int IdxNum = getNumElements(), typename = EnableIfOneIndex<IdxNum>>
-  SwizzleOp &operator=(DataT &&Rhs) {
+  template <int IdxNum = getNumElements(), typename RelayVecT = VecT>
+  std::enable_if_t<HasOneIndex<IdxNum> && VecIsMutable<RelayVecT>, SwizzleOp> &
+  operator=(DataT &&Rhs) {
     std::array<int, IdxNum> Idxs{Indexes...};
     m_Vector->setValue(Idxs[0], Rhs);
     return *this;
@@ -2219,10 +2234,10 @@ class SwizzleOp {
     return NewLHOp<RhsOperation, LShift, Indexes...>(m_Vector, *this, Rhs);
   }
 
-  template <
-      typename T1, typename T2, typename T3, template <typename> class T4,
-      int... T5,
-      typename = typename std::enable_if_t<sizeof...(T5) == getNumElements()>>
+  template <typename T1, typename T2, typename T3, template <typename> class T4,
+            int... T5,
+            typename = typename std::enable_if_t<
+                sizeof...(T5) == getNumElements() && VecIsMutable<>>>
   SwizzleOp &operator=(const SwizzleOp<T1, T2, T3, T4, T5...> &Rhs) {
     std::array<int, getNumElements()> Idxs{Indexes...};
     for (size_t I = 0; I < Idxs.size(); ++I) {
@@ -2231,10 +2246,10 @@ class SwizzleOp {
     return *this;
   }
 
-  template <
-      typename T1, typename T2, typename T3, template <typename> class T4,
-      int... T5,
-      typename = typename std::enable_if_t<sizeof...(T5) == getNumElements()>>
+  template <typename T1, typename T2, typename T3, template <typename> class T4,
+            int... T5,
+            typename = typename std::enable_if_t<
+                sizeof...(T5) == getNumElements() && VecIsMutable<>>>
   SwizzleOp &operator=(SwizzleOp<T1, T2, T3, T4, T5...> &&Rhs) {
     std::array<int, getNumElements()> Idxs{Indexes...};
     for (size_t I = 0; I < Idxs.size(); ++I) {
diff --git a/sycl/test/basic_tests/vectors/const_swizzle_negative.cpp b/sycl/test/basic_tests/vectors/const_swizzle_negative.cpp
new file mode 100644
index 0000000000000..c6f756132af92
--- /dev/null
+++ b/sycl/test/basic_tests/vectors/const_swizzle_negative.cpp
@@ -0,0 +1,91 @@
+// REQUIRES: preview-breaking-changes-supported
+// RUN: %clangxx -fsycl-device-only -ferror-limit=0 -Xclang -fsycl-is-device -fsyntax-only -fpreview-breaking-changes -Xclang -verify -Xclang -verify-ignore-unexpected=note %s
+
+#include <sycl/sycl.hpp>
+
+int main() {
+  sycl::queue Q;
+  Q.single_task([]() {
+    const sycl::vec<int, 4> X{1};
+
+    // expected-error@+1 {{no viable overloaded '='}}
+    X.swizzle<0>() = 1;
+    // expected-error@+1 {{no viable overloaded '='}}
+    X.swizzle<0>() = sycl::vec<int, 1>{1};
+    // expected-error@+1 {{no viable overloaded '='}}
+    X.swizzle<0, 2>() = sycl::vec<int, 2>{1};
+
+    // expected-error@+1 {{no viable overloaded '+='}}
+    X.swizzle<0>() += 1;
+    // expected-error@+1 {{no viable overloaded '-='}}
+    X.swizzle<0>() -= 1;
+    // expected-error@+1 {{no viable overloaded '*='}}
+    X.swizzle<0>() *= 1;
+    // expected-error@+1 {{no viable overloaded '/='}}
+    X.swizzle<0>() /= 1;
+    // expected-error@+1 {{no viable overloaded '%='}}
+    X.swizzle<0>() %= 1;
+    // expected-error@+1 {{no viable overloaded '&='}}
+    X.swizzle<0>() &= 1;
+    // expected-error@+1 {{no viable overloaded '|='}}
+    X.swizzle<0>() |= 1;
+    // expected-error@+1 {{no viable overloaded '^='}}
+    X.swizzle<0>() ^= 1;
+    // expected-error@+1 {{no viable overloaded '>>='}}
+    X.swizzle<0>() >>= 1;
+    // expected-error@+1 {{no viable overloaded '<<='}}
+    X.swizzle<0>() <<= 1;
+
+    // expected-error@+1 {{no viable overloaded '+='}}
+    X.swizzle<0>() += sycl::vec<int, 1>{1};
+    // expected-error@+1 {{no viable overloaded '-='}}
+    X.swizzle<0>() -= sycl::vec<int, 1>{1};
+    // expected-error@+1 {{no viable overloaded '*='}}
+    X.swizzle<0>() *= sycl::vec<int, 1>{1};
+    // expected-error@+1 {{no viable overloaded '/='}}
+    X.swizzle<0>() /= sycl::vec<int, 1>{1};
+    // expected-error@+1 {{no viable overloaded '%='}}
+    X.swizzle<0>() %= sycl::vec<int, 1>{1};
+    // expected-error@+1 {{no viable overloaded '&='}}
+    X.swizzle<0>() &= sycl::vec<int, 1>{1};
+    // expected-error@+1 {{no viable overloaded '|='}}
+    X.swizzle<0>() |= sycl::vec<int, 1>{1};
+    // expected-error@+1 {{no viable overloaded '^='}}
+    X.swizzle<0>() ^= sycl::vec<int, 1>{1};
+    // expected-error@+1 {{no viable overloaded '>>='}}
+    X.swizzle<0>() >>= sycl::vec<int, 1>{1};
+    // expected-error@+1 {{no viable overloaded '<<='}}
+    X.swizzle<0>() <<= sycl::vec<int, 1>{1};
+
+    // expected-error@+1 {{no viable overloaded '+='}}
+    X.swizzle<0>() += X.swizzle<1>();
+    // expected-error@+1 {{no viable overloaded '-='}}
+    X.swizzle<0>() -= X.swizzle<1>();
+    // expected-error@+1 {{no viable overloaded '*='}}
+    X.swizzle<0>() *= X.swizzle<1>();
+    // expected-error@+1 {{no viable overloaded '/='}}
+    X.swizzle<0>() /= X.swizzle<1>();
+    // expected-error@+1 {{no viable overloaded '%='}}
+    X.swizzle<0>() %= X.swizzle<1>();
+    // expected-error@+1 {{no viable overloaded '&='}}
+    X.swizzle<0>() &= X.swizzle<1>();
+    // expected-error@+1 {{no viable overloaded '|='}}
+    X.swizzle<0>() |= X.swizzle<1>();
+    // expected-error@+1 {{no viable overloaded '^='}}
+    X.swizzle<0>() ^= X.swizzle<1>();
+    // expected-error@+1 {{no viable overloaded '>>='}}
+    X.swizzle<0>() >>= X.swizzle<1>();
+    // expected-error@+1 {{no viable overloaded '<<='}}
+    X.swizzle<0>() <<= X.swizzle<1>();
+
+    // expected-error@+1 {{cannot increment value of type}}
+    X.swizzle<0>()++;
+    // expected-error@+1 {{cannot increment value of type}}
+    ++X.swizzle<0>();
+    // expected-error@+1 {{cannot decrement value of type}}
+    X.swizzle<0>()--;
+    // expected-error@+1 {{cannot decrement value of type}}
+    --X.swizzle<0>();
+  });
+  return 0;
+}

From 860ad1f77b391c8463f39efc7441befb8a57ab59 Mon Sep 17 00:00:00 2001
From: "Larsen, Steffen" <steffen.larsen@intel.com>
Date: Wed, 20 Mar 2024 02:43:45 -0700
Subject: [PATCH 3/6] Also disallow load for const vecs and make it const

Signed-off-by: Larsen, Steffen <steffen.larsen@intel.com>
---
 sycl/include/sycl/types.hpp                              | 6 ++++--
 sycl/test/basic_tests/vectors/const_swizzle_negative.cpp | 6 ++++++
 2 files changed, 10 insertions(+), 2 deletions(-)

diff --git a/sycl/include/sycl/types.hpp b/sycl/include/sycl/types.hpp
index e3be3ed524d84..4a3bc2a39ee1b 100644
--- a/sycl/include/sycl/types.hpp
+++ b/sycl/include/sycl/types.hpp
@@ -2391,8 +2391,10 @@ class SwizzleOp {
 
   // Leave store() interface to automatic conversion to vec<>.
   // Load to vec_t and then assign to swizzle.
-  template <access::address_space Space, access::decorated DecorateAddress>
-  void load(size_t offset, multi_ptr<DataT, Space, DecorateAddress> ptr) {
+  template <access::address_space Space, access::decorated DecorateAddress,
+            typename RelayVecT = VecT>
+  std::enable_if_t<VecIsMutable<RelayVecT>, void>
+  load(size_t offset, multi_ptr<DataT, Space, DecorateAddress> ptr) const {
     vec_t Tmp;
     Tmp.template load(offset, ptr);
     *this = Tmp;
diff --git a/sycl/test/basic_tests/vectors/const_swizzle_negative.cpp b/sycl/test/basic_tests/vectors/const_swizzle_negative.cpp
index c6f756132af92..1d14994a4d2b8 100644
--- a/sycl/test/basic_tests/vectors/const_swizzle_negative.cpp
+++ b/sycl/test/basic_tests/vectors/const_swizzle_negative.cpp
@@ -86,6 +86,12 @@ int main() {
     X.swizzle<0>()--;
     // expected-error@+1 {{cannot decrement value of type}}
     --X.swizzle<0>();
+
+    int I = 1;
+    // expected-error@+1 {{no matching member function for call to 'load'}}
+    X.load(0,
+           sycl::address_space_cast<sycl::access::address_space::private_space,
+                                    sycl::access::decorated::no>(&I));
   });
   return 0;
 }

From 331f1d0ab2efa62299355ff8ad903cd10b184c06 Mon Sep 17 00:00:00 2001
From: "Larsen, Steffen" <steffen.larsen@intel.com>
Date: Mon, 13 May 2024 22:54:15 -0700
Subject: [PATCH 4/6] Use swizzle in load

Signed-off-by: Larsen, Steffen <steffen.larsen@intel.com>
---
 sycl/test/basic_tests/vectors/const_swizzle_negative.cpp | 4 ++--
 1 file changed, 2 insertions(+), 2 deletions(-)

diff --git a/sycl/test/basic_tests/vectors/const_swizzle_negative.cpp b/sycl/test/basic_tests/vectors/const_swizzle_negative.cpp
index 1d14994a4d2b8..ca41dd78b5fbb 100644
--- a/sycl/test/basic_tests/vectors/const_swizzle_negative.cpp
+++ b/sycl/test/basic_tests/vectors/const_swizzle_negative.cpp
@@ -89,8 +89,8 @@ int main() {
 
     int I = 1;
     // expected-error@+1 {{no matching member function for call to 'load'}}
-    X.load(0,
-           sycl::address_space_cast<sycl::access::address_space::private_space,
+    X.swizzle<0>().load(
+        0, sycl::address_space_cast<sycl::access::address_space::private_space,
                                     sycl::access::decorated::no>(&I));
   });
   return 0;

From 396de3c3c555b7102bbfd2294f621fc24a28b2d6 Mon Sep 17 00:00:00 2001
From: "Larsen, Steffen" <steffen.larsen@intel.com>
Date: Wed, 26 Jun 2024 00:18:13 -0700
Subject: [PATCH 5/6] Fix tests

Signed-off-by: Larsen, Steffen <steffen.larsen@intel.com>
---
 sycl/test-e2e/Regression/swizzle_opassign.cpp            | 3 ++-
 sycl/test/basic_tests/vectors/const_swizzle_negative.cpp | 3 ++-
 2 files changed, 4 insertions(+), 2 deletions(-)

diff --git a/sycl/test-e2e/Regression/swizzle_opassign.cpp b/sycl/test-e2e/Regression/swizzle_opassign.cpp
index 3a6a2357ffdc4..83521f74217c2 100644
--- a/sycl/test-e2e/Regression/swizzle_opassign.cpp
+++ b/sycl/test-e2e/Regression/swizzle_opassign.cpp
@@ -7,7 +7,8 @@
 // Tests that the mutating operators (+=, -=, ..., ++, --) on swizzles compile
 // and correctly mutate the elements in the corresponding vector.
 
-#include <sycl/sycl.hpp>
+#include <sycl/detail/core.hpp>
+#include <sycl/types.hpp>
 
 constexpr size_t NumOps = 14;
 constexpr std::string_view OpNames[NumOps] = {
diff --git a/sycl/test/basic_tests/vectors/const_swizzle_negative.cpp b/sycl/test/basic_tests/vectors/const_swizzle_negative.cpp
index ca41dd78b5fbb..ac1629bd6d1db 100644
--- a/sycl/test/basic_tests/vectors/const_swizzle_negative.cpp
+++ b/sycl/test/basic_tests/vectors/const_swizzle_negative.cpp
@@ -1,7 +1,8 @@
 // REQUIRES: preview-breaking-changes-supported
 // RUN: %clangxx -fsycl-device-only -ferror-limit=0 -Xclang -fsycl-is-device -fsyntax-only -fpreview-breaking-changes -Xclang -verify -Xclang -verify-ignore-unexpected=note %s
 
-#include <sycl/sycl.hpp>
+#include <sycl/detail/core.hpp>
+#include <sycl/types.hpp>
 
 int main() {
   sycl::queue Q;

From 955c841795d139b30a1837e8e4ee00277dd9a113 Mon Sep 17 00:00:00 2001
From: "Larsen, Steffen" <steffen.larsen@intel.com>
Date: Wed, 26 Jun 2024 01:42:48 -0700
Subject: [PATCH 6/6] Add missing include to test

Signed-off-by: Larsen, Steffen <steffen.larsen@intel.com>
---
 sycl/test-e2e/Regression/swizzle_opassign.cpp | 1 +
 1 file changed, 1 insertion(+)

diff --git a/sycl/test-e2e/Regression/swizzle_opassign.cpp b/sycl/test-e2e/Regression/swizzle_opassign.cpp
index 83521f74217c2..99c163fe5c98b 100644
--- a/sycl/test-e2e/Regression/swizzle_opassign.cpp
+++ b/sycl/test-e2e/Regression/swizzle_opassign.cpp
@@ -9,6 +9,7 @@
 
 #include <sycl/detail/core.hpp>
 #include <sycl/types.hpp>
+#include <sycl/usm.hpp>
 
 constexpr size_t NumOps = 14;
 constexpr std::string_view OpNames[NumOps] = {