From 1f463fed4cb43b8bf1e988e3ae7e23584792f099 Mon Sep 17 00:00:00 2001 From: Dmitri Mokhov Date: Mon, 25 Jan 2021 21:48:14 -0800 Subject: [PATCH 1/3] [SYCL] Fix long long support in vec::convert on Windows - Keep `long long` and `unsigned long long` in `is_standard_type`. Excluding them resulted in duplicate `convertImpl` definitions on Windows when converting from `long long` or `unsigned long long`: SPIR-V-based and the non-standard type version. - For to-int SPIR-V-based `convertImpl` overload resolution, compare the OpenCL return type with cl_DestType instead of simply DestType ((u)char, (u)short, (u)int, (u)long). When converting to `long long` or `unsigned long long` on Windows, the non-standard type version was used instead of these functions, because `cl_long == long long != long` there. --- sycl/include/CL/sycl/types.hpp | 11 +++++------ 1 file changed, 5 insertions(+), 6 deletions(-) diff --git a/sycl/include/CL/sycl/types.hpp b/sycl/include/CL/sycl/types.hpp index a5b07eb29153..fb641df6ed5a 100644 --- a/sycl/include/CL/sycl/types.hpp +++ b/sycl/include/CL/sycl/types.hpp @@ -242,9 +242,8 @@ using is_float_to_float = std::integral_constant::value && detail::is_floating_point::value>; template -using is_standard_type = std::integral_constant< - bool, detail::is_sgentype::value && !std::is_same::value && - !std::is_same::value>; +using is_standard_type = + std::integral_constant::value>; template @@ -330,7 +329,7 @@ convertImpl(T Value) { typename OpenCLT, typename OpenCLR> \ detail::enable_if_t::value && \ !std::is_same::value && \ - (std::is_same::value || \ + (std::is_same::value || \ (std::is_same::value && \ std::is_same::value)), \ R> \ @@ -352,7 +351,7 @@ __SYCL_GENERATE_CONVERT_IMPL(long) typename OpenCLT, typename OpenCLR> \ detail::enable_if_t::value && \ !std::is_same::value && \ - std::is_same::value, \ + std::is_same::value, \ R> \ convertImpl(T Value) { \ OpenCLT OpValue = cl::sycl::detail::convertDataToType(Value); \ @@ -454,7 +453,7 @@ __SYCL_GENERATE_CONVERT_IMPL_FOR_ROUNDING_MODE(rtn, Rtn) template \ detail::enable_if_t::value && \ - (std::is_same::value || \ + (std::is_same::value || \ std::is_same::value && \ std::is_same::value) && \ RoundingModeCondition::value, \ From b777b39d770f02a367bba13262da6f048976ac8d Mon Sep 17 00:00:00 2001 From: Dmitri Mokhov Date: Mon, 1 Feb 2021 12:45:24 -0600 Subject: [PATCH 2/3] Add a test for various type combinations. --- sycl/test/basic_tests/vectors/vectors.cpp | 52 +++++++++++++++++++---- 1 file changed, 44 insertions(+), 8 deletions(-) diff --git a/sycl/test/basic_tests/vectors/vectors.cpp b/sycl/test/basic_tests/vectors/vectors.cpp index dafbbc20a3e0..aa48aace9df8 100644 --- a/sycl/test/basic_tests/vectors/vectors.cpp +++ b/sycl/test/basic_tests/vectors/vectors.cpp @@ -10,21 +10,48 @@ #define SYCL_SIMPLE_SWIZZLES #include -using namespace cl::sycl; -void check_vectors(int4 a, int4 b, int4 c, int4 gold) { - int4 result = a * (int4)b.y() + c; +void check_vectors(cl::sycl::int4 a, cl::sycl::int4 b, cl::sycl::int4 c, + cl::sycl::int4 gold) { + cl::sycl::int4 result = a * (cl::sycl::int4)b.y() + c; assert((int)result.x() == (int)gold.x()); assert((int)result.y() == (int)gold.y()); - assert((int)result.w() == (int)gold.w()); assert((int)result.z() == (int)gold.z()); + assert((int)result.w() == (int)gold.w()); +} + +template void check_convert() { + cl::sycl::vec vec{1, 2, 3, 4}; + cl::sycl::vec result = vec.template convert(); + assert((int)result.x() == (int)vec.x()); + assert((int)result.y() == (int)vec.y()); + assert((int)result.z() == (int)vec.z()); + assert((int)result.w() == (int)vec.w()); +} + +template void check_signed_unsigned_convert_to() { + check_convert(); + check_convert>(); + check_convert, To>(); + check_convert, + cl::sycl::detail::make_unsigned_t>(); +} + +template void check_convert_from() { + check_signed_unsigned_convert_to(); + check_signed_unsigned_convert_to(); + check_signed_unsigned_convert_to(); + check_signed_unsigned_convert_to(); + check_signed_unsigned_convert_to(); + check_signed_unsigned_convert_to(); + check_signed_unsigned_convert_to(); } int main() { - int4 a = {1, 2, 3, 4}; - const int4 b = {10, 20, 30, 40}; - const int4 gold = {21, 42, 90, 120}; - const int2 a_xy = a.xy(); + cl::sycl::int4 a = {1, 2, 3, 4}; + const cl::sycl::int4 b = {10, 20, 30, 40}; + const cl::sycl::int4 gold = {21, 42, 90, 120}; + const cl::sycl::int2 a_xy = a.xy(); check_vectors(a, b, {1, 2, 30, 40}, gold); check_vectors(a, b, {a.x(), a.y(), b.z(), b.w()}, gold); check_vectors(a, b, {a.x(), 2, b.z(), 40}, gold); @@ -82,5 +109,14 @@ int main() { assert((std::is_same, cl::sycl::ulong8>::value)); assert((std::is_same, cl::sycl::ulong16>::value)); + // Check convert() from and to various types. + check_convert_from(); + check_convert_from(); + check_convert_from(); + check_convert_from(); + check_convert_from(); + check_convert_from(); + check_convert_from(); + return 0; } From 0cd20fe7213775afc5cb74f929a7637122649e75 Mon Sep 17 00:00:00 2001 From: Dmitri Mokhov Date: Tue, 2 Feb 2021 11:06:49 -0600 Subject: [PATCH 3/3] Added char, short, int, long, long long conversion tests and dropped cl:: prefix. --- sycl/test/basic_tests/vectors/vectors.cpp | 76 +++++++++++++---------- 1 file changed, 42 insertions(+), 34 deletions(-) diff --git a/sycl/test/basic_tests/vectors/vectors.cpp b/sycl/test/basic_tests/vectors/vectors.cpp index aa48aace9df8..a3175f366d10 100644 --- a/sycl/test/basic_tests/vectors/vectors.cpp +++ b/sycl/test/basic_tests/vectors/vectors.cpp @@ -11,9 +11,8 @@ #define SYCL_SIMPLE_SWIZZLES #include -void check_vectors(cl::sycl::int4 a, cl::sycl::int4 b, cl::sycl::int4 c, - cl::sycl::int4 gold) { - cl::sycl::int4 result = a * (cl::sycl::int4)b.y() + c; +void check_vectors(sycl::int4 a, sycl::int4 b, sycl::int4 c, sycl::int4 gold) { + sycl::int4 result = a * (sycl::int4)b.y() + c; assert((int)result.x() == (int)gold.x()); assert((int)result.y() == (int)gold.y()); assert((int)result.z() == (int)gold.z()); @@ -21,8 +20,8 @@ void check_vectors(cl::sycl::int4 a, cl::sycl::int4 b, cl::sycl::int4 c, } template void check_convert() { - cl::sycl::vec vec{1, 2, 3, 4}; - cl::sycl::vec result = vec.template convert(); + sycl::vec vec{1, 2, 3, 4}; + sycl::vec result = vec.template convert(); assert((int)result.x() == (int)vec.x()); assert((int)result.y() == (int)vec.y()); assert((int)result.z() == (int)vec.z()); @@ -31,10 +30,10 @@ template void check_convert() { template void check_signed_unsigned_convert_to() { check_convert(); - check_convert>(); - check_convert, To>(); - check_convert, - cl::sycl::detail::make_unsigned_t>(); + check_convert>(); + check_convert, To>(); + check_convert, + sycl::detail::make_unsigned_t>(); } template void check_convert_from() { @@ -42,16 +41,21 @@ template void check_convert_from() { check_signed_unsigned_convert_to(); check_signed_unsigned_convert_to(); check_signed_unsigned_convert_to(); + check_signed_unsigned_convert_to(); + check_signed_unsigned_convert_to(); + check_signed_unsigned_convert_to(); + check_signed_unsigned_convert_to(); + check_signed_unsigned_convert_to(); check_signed_unsigned_convert_to(); check_signed_unsigned_convert_to(); check_signed_unsigned_convert_to(); } int main() { - cl::sycl::int4 a = {1, 2, 3, 4}; - const cl::sycl::int4 b = {10, 20, 30, 40}; - const cl::sycl::int4 gold = {21, 42, 90, 120}; - const cl::sycl::int2 a_xy = a.xy(); + sycl::int4 a = {1, 2, 3, 4}; + const sycl::int4 b = {10, 20, 30, 40}; + const sycl::int4 gold = {21, 42, 90, 120}; + const sycl::int2 a_xy = a.xy(); check_vectors(a, b, {1, 2, 30, 40}, gold); check_vectors(a, b, {a.x(), a.y(), b.z(), b.w()}, gold); check_vectors(a, b, {a.x(), 2, b.z(), 40}, gold); @@ -60,11 +64,11 @@ int main() { check_vectors(a, b, {a.xy(), b.zw()}, gold); // Constructing vector from a scalar - cl::sycl::vec vec_from_one_elem(1); + sycl::vec vec_from_one_elem(1); // implicit conversion - cl::sycl::vec vec_2(1, 2); - cl::sycl::vec vec_4(0, vec_2, 3); + sycl::vec vec_2(1, 2); + sycl::vec vec_4(0, vec_2, 3); assert(vec_4.get_count() == 4); assert(static_cast(vec_4.x()) == static_cast(0)); @@ -74,10 +78,10 @@ int main() { // explicit conversion int64_t(vec_2.x()); - cl::sycl::int4(vec_2.x()); + sycl::int4(vec_2.x()); // Check broadcasting operator= - cl::sycl::vec b_vec(1.0); + sycl::vec b_vec(1.0); b_vec = 0.5; assert(static_cast(b_vec.x()) == static_cast(0.5)); assert(static_cast(b_vec.y()) == static_cast(0.5)); @@ -87,33 +91,37 @@ int main() { // Check that vector with 'unsigned long long' elements has enough bits to // store value. unsigned long long ull_ref = 1ull - 2ull; - auto ull_vec = cl::sycl::vec(ull_ref); - unsigned long long ull_val = ull_vec.template swizzle(); + auto ull_vec = sycl::vec(ull_ref); + unsigned long long ull_val = ull_vec.template swizzle(); assert(ull_val == ull_ref); // Check that the function as() in swizzle vec class is working correctly - cl::sycl::vec inputVec = cl::sycl::vec(0, 1); - auto asVec = - inputVec.template swizzle() - .template as>(); + sycl::vec inputVec = sycl::vec(0, 1); + auto asVec = inputVec.template swizzle() + .template as>(); // Check that [u]long[n] type aliases match vec<[unsigned] long, n> types. - assert((std::is_same, cl::sycl::long2>::value)); - assert((std::is_same, cl::sycl::long3>::value)); - assert((std::is_same, cl::sycl::long4>::value)); - assert((std::is_same, cl::sycl::long8>::value)); - assert((std::is_same, cl::sycl::long16>::value)); - assert((std::is_same, cl::sycl::ulong2>::value)); - assert((std::is_same, cl::sycl::ulong3>::value)); - assert((std::is_same, cl::sycl::ulong4>::value)); - assert((std::is_same, cl::sycl::ulong8>::value)); - assert((std::is_same, cl::sycl::ulong16>::value)); + assert((std::is_same, sycl::long2>::value)); + assert((std::is_same, sycl::long3>::value)); + assert((std::is_same, sycl::long4>::value)); + assert((std::is_same, sycl::long8>::value)); + assert((std::is_same, sycl::long16>::value)); + assert((std::is_same, sycl::ulong2>::value)); + assert((std::is_same, sycl::ulong3>::value)); + assert((std::is_same, sycl::ulong4>::value)); + assert((std::is_same, sycl::ulong8>::value)); + assert((std::is_same, sycl::ulong16>::value)); // Check convert() from and to various types. check_convert_from(); check_convert_from(); check_convert_from(); check_convert_from(); + check_convert_from(); + check_convert_from(); + check_convert_from(); + check_convert_from(); + check_convert_from(); check_convert_from(); check_convert_from(); check_convert_from();