diff --git a/sycl/include/sycl/types.hpp b/sycl/include/sycl/types.hpp index 426a1868705aa..b9ec19748e05f 100644 --- a/sycl/include/sycl/types.hpp +++ b/sycl/include/sycl/types.hpp @@ -1544,8 +1544,37 @@ template class OperationCurrentT, int... Indexes> class SwizzleOp { using DataT = typename VecT::element_type; - using CommonDataT = std::common_type_t; + // Certain operators return a vector with a different element type. Also, the + // left and right operand types may differ. CommonDataT selects a result type + // based on these types to ensure that the result value can be represented. + // + // Example 1: + // sycl::vec vec{...}; + // auto result = 300u + vec.x(); + // + // CommonDataT is std::common_type_t since + // it's larger than unsigned char. + // + // Example 2: + // sycl::vec vec{...}; + // auto result = vec.template swizzle() && vec; + // + // CommonDataT is DataT since operator&& returns a vector with element type + // int8_t, which is larger than bool. + // + // Example 3: + // sycl::vec vec{...}; auto swlo = vec.lo(); + // auto result = swlo == swlo; + // + // CommonDataT is DataT since operator== returns a vector with element type + // int8_t, which is the same size as std::byte. std::common_type_t + // can't be used here since there's no type that int8_t and std::byte can both + // be implicitly converted to. + using OpLeftDataT = typename OperationLeftT::DataT; + using OpRightDataT = typename OperationRightT::DataT; + using CommonDataT = std::conditional_t< + sizeof(DataT) >= sizeof(std::common_type_t), + DataT, std::common_type_t>; static constexpr int getNumElements() { return sizeof...(Indexes); } using rel_t = detail::rel_t; diff --git a/sycl/test-e2e/Regression/vec_rel_swizzle_ops.cpp b/sycl/test-e2e/Regression/vec_rel_swizzle_ops.cpp new file mode 100644 index 0000000000000..36d4443824db7 --- /dev/null +++ b/sycl/test-e2e/Regression/vec_rel_swizzle_ops.cpp @@ -0,0 +1,68 @@ +// RUN: %{build} -o %t.out +// RUN: %{run} %t.out + +// RUN: %if preview-breaking-changes-supported %{ %clangxx -fsycl -fpreview-breaking-changes %s -o %t2.out %} +// RUN: %if preview-breaking-changes-supported %{ %{run} %t2.out %} + +#include +#include + +template +bool testAndOperator(const std::string &typeName) { + constexpr int N = 5; + std::array results{}; + + sycl::queue q; + sycl::buffer buffer{results.data(), N}; + q.submit([&](sycl::handler &cgh) { + sycl::accessor acc{buffer, cgh, sycl::write_only}; + cgh.parallel_for(sycl::range<1>{1}, [=](sycl::id<1> id) { + auto testVec1 = sycl::vec(static_cast(1)); + auto testVec2 = sycl::vec(static_cast(2)); + sycl::vec resVec; + + ResultT expected = static_cast( + -(static_cast(1) && static_cast(2))); + acc[0] = expected; + + // LHS swizzle + resVec = testVec1.template swizzle() && testVec2; + acc[1] = resVec[0]; + + // RHS swizzle + resVec = testVec1 && testVec2.template swizzle(); + acc[2] = resVec[0]; + + // No swizzle + resVec = testVec1 && testVec2; + acc[3] = resVec[0]; + + // Both swizzle + resVec = testVec1.template swizzle() && + testVec2.template swizzle(); + acc[4] = resVec[0]; + }); + }).wait(); + + bool passed = true; + ResultT expected = results[0]; + + std::cout << "Testing with T = " << typeName << std::endl; + std::cout << "Expected: " << (int)expected << std::endl; + for (int i = 1; i < N; i++) { + std::cout << "Test " << (i - 1) << ": " << ((int)results[i]) << std::endl; + passed &= expected == results[i]; + } + std::cout << std::endl; + return passed; +} + +int main() { + bool passed = true; + passed &= testAndOperator("bool"); + passed &= testAndOperator("std::int8_t"); + passed &= testAndOperator("float"); + passed &= testAndOperator("int"); + std::cout << (passed ? "Pass" : "Fail") << std::endl; + return (passed ? EXIT_SUCCESS : EXIT_FAILURE); +}