[ABI-Break][SYCL] Switch sycl::vec to preview unconditionally (#13182)

Author: aelovikov-intel

sycl/include/sycl/detail/generic_type_traits.hpp

@@ -429,8 +429,12 @@ template <typename To, typename From> auto convertFromOpenCLTypeFor(From &&x) {
     using OpenCLType = decltype(convertToOpenCLType(std::declval<To>()));
     static_assert(std::is_same_v<std::remove_reference_t<From>, OpenCLType>);
     static_assert(sizeof(OpenCLType) == sizeof(To));
-    if constexpr (is_vec_v<To> && is_vec_v<From>)
-      return x.template as<To>();
+    using To_noref = std::remove_reference_t<To>;
+    using From_noref = std::remove_reference_t<From>;
+    if constexpr (is_vec_v<To_noref> && is_vec_v<From_noref>)
+      return x.template as<To_noref>();
+    else if constexpr (is_vec_v<To_noref> && is_ext_vector_v<From_noref>)
+      return To_noref{bit_cast<typename To_noref::vector_t>(x)};
     else
       return static_cast<To>(x);
   }

sycl/include/sycl/detail/vector_traits.hpp

@@ -16,7 +16,6 @@ inline namespace _V1 {
 namespace detail {
 
 // 4.10.2.6 Memory layout and alignment
-#ifdef __INTEL_PREVIEW_BREAKING_CHANGES
 // due to MSVC the maximum alignment for sycl::vec is 64 and this proposed
 // change is being brought to the spec committee.
 constexpr size_t MaxVecAlignment = 64;
@@ -29,12 +28,6 @@ struct vector_alignment_impl
           std::integral_constant<size_t,
                                  (std::min)(sizeof(T) * N, MaxVecAlignment)>> {
 };
-#else
-template <typename T, size_t N>
-struct vector_alignment_impl
-    : std::conditional_t<N == 3, std::integral_constant<int, sizeof(T) * 4>,
-                         std::integral_constant<int, sizeof(T) * N>> {};
-#endif
 
 template <typename T, size_t N>
 struct vector_alignment

sycl/include/sycl/half_type.hpp

@@ -259,36 +259,11 @@ using BIsRepresentationT = half;
 // for vec because they are actually defined as an integer type under the
 // hood. As a result half values will be converted to the integer and passed
 // as a kernel argument which is expected to be floating point number.
-#ifdef __INTEL_PREVIEW_BREAKING_CHANGES
 using Vec2StorageT = std::array<StorageT, 2>;
 using Vec3StorageT = std::array<StorageT, 3>;
 using Vec4StorageT = std::array<StorageT, 4>;
 using Vec8StorageT = std::array<StorageT, 8>;
 using Vec16StorageT = std::array<StorageT, 16>;
-#else  // __INTEL_PREVIEW_BREAKING_CHANGES
-template <int NumElements> struct half_vec {
-  alignas(
-      vector_alignment<StorageT, NumElements>::value) StorageT s[NumElements];
-
-  __SYCL_CONSTEXPR_HALF half_vec() : s{0.0f} { initialize_data(); }
-  template <typename... Ts,
-            typename = std::enable_if_t<(sizeof...(Ts) == NumElements) &&
-                                        (std::is_same_v<half, Ts> && ...)>>
-  __SYCL_CONSTEXPR_HALF half_vec(const Ts &...hs) : s{hs...} {}
-
-  constexpr void initialize_data() {
-    for (size_t i = 0; i < NumElements; ++i) {
-      s[i] = StorageT(0.0f);
-    }
-  }
-};
-
-using Vec2StorageT = half_vec<2>;
-using Vec3StorageT = half_vec<3>;
-using Vec4StorageT = half_vec<4>;
-using Vec8StorageT = half_vec<8>;
-using Vec16StorageT = half_vec<16>;
-#endif // __INTEL_PREVIEW_BREAKING_CHANGES
 #endif // SYCL_DEVICE_ONLY
 
 #ifndef __SYCL_DEVICE_ONLY__