Skip to content

Commit 76eab93

Browse files
fadeevalfadeeval
authored
[SYCL] Fix conversion of long long vector types (#1734)
This is a regression introduced by 096d0a0. Signed-off-by: Aleksander Fadeev <aleksander.fadeev@intel.com>
1 parent c9fd72c commit 76eab93

File tree

2 files changed

+38
-23
lines changed

2 files changed

+38
-23
lines changed

‎sycl/include/CL/sycl/types.hpp

Lines changed: 29 additions & 23 deletions
Original file line numberDiff line numberDiff line change
@@ -314,14 +314,11 @@ using Rtp = detail::bool_constant<Mode == rounding_mode::rtp>;
314314
template <rounding_mode Mode>
315315
using Rtn = detail::bool_constant<Mode == rounding_mode::rtn>;
316316

317-
// convert signed and unsigned types with an equal size and diff names
317+
// convert types with an equal size and diff names
318318
template <typename T, typename R, rounding_mode roundingMode, typename OpenCLT,
319319
typename OpenCLR>
320-
detail::enable_if_t<!std::is_same<T, R>::value &&
321-
(is_sint_to_sint<T, R>::value ||
322-
is_uint_to_uint<T, R>::value) &&
323-
std::is_same<OpenCLT, OpenCLR>::value,
324-
R>
320+
detail::enable_if_t<
321+
!std::is_same<T, R>::value && std::is_same<OpenCLT, OpenCLR>::value, R>
325322
convertImpl(T Value) {
326323
return static_cast<R>(Value);
327324
}
@@ -330,12 +327,11 @@ convertImpl(T Value) {
330327
#define __SYCL_GENERATE_CONVERT_IMPL(DestType) \
331328
template <typename T, typename R, rounding_mode roundingMode, \
332329
typename OpenCLT, typename OpenCLR> \
333-
detail::enable_if_t<!std::is_same<T, R>::value && \
334-
is_sint_to_sint<T, R>::value && \
330+
detail::enable_if_t<is_sint_to_sint<T, R>::value && \
331+
!std::is_same<OpenCLT, OpenCLR>::value && \
335332
(std::is_same<OpenCLR, DestType>::value || \
336333
std::is_same<OpenCLR, signed char>::value && \
337-
std::is_same<DestType, char>::value) && \
338-
!std::is_same<OpenCLT, OpenCLR>::value, \
334+
std::is_same<DestType, char>::value), \
339335
R> \
340336
convertImpl(T Value) { \
341337
OpenCLT OpValue = cl::sycl::detail::convertDataToType<T, OpenCLT>(Value); \
@@ -346,18 +342,16 @@ __SYCL_GENERATE_CONVERT_IMPL(char)
346342
__SYCL_GENERATE_CONVERT_IMPL(short)
347343
__SYCL_GENERATE_CONVERT_IMPL(int)
348344
__SYCL_GENERATE_CONVERT_IMPL(long)
349-
__SYCL_GENERATE_CONVERT_IMPL(longlong)
350345

351346
#undef __SYCL_GENERATE_CONVERT_IMPL
352347

353348
// unsigned to unsigned
354349
#define __SYCL_GENERATE_CONVERT_IMPL(DestType) \
355350
template <typename T, typename R, rounding_mode roundingMode, \
356351
typename OpenCLT, typename OpenCLR> \
357-
detail::enable_if_t<!std::is_same<T, R>::value && \
358-
is_uint_to_uint<T, R>::value && \
359-
std::is_same<OpenCLR, DestType>::value && \
360-
!std::is_same<OpenCLT, OpenCLR>::value, \
352+
detail::enable_if_t<is_uint_to_uint<T, R>::value && \
353+
!std::is_same<OpenCLT, OpenCLR>::value && \
354+
std::is_same<OpenCLR, DestType>::value, \
361355
R> \
362356
convertImpl(T Value) { \
363357
OpenCLT OpValue = cl::sycl::detail::convertDataToType<T, OpenCLT>(Value); \
@@ -374,16 +368,23 @@ __SYCL_GENERATE_CONVERT_IMPL(ulong)
374368
// unsigned to (from) signed
375369
template <typename T, typename R, rounding_mode roundingMode, typename OpenCLT,
376370
typename OpenCLR>
377-
detail::enable_if_t<is_sint_to_from_uint<T, R>::value, R> convertImpl(T Value) {
371+
detail::enable_if_t<is_sint_to_from_uint<T, R>::value &&
372+
is_standard_type<OpenCLT>::value &&
373+
is_standard_type<OpenCLR>::value,
374+
R>
375+
convertImpl(T Value) {
378376
return static_cast<R>(Value);
379377
}
380378

381379
// sint to float
382380
#define __SYCL_GENERATE_CONVERT_IMPL(SPIRVOp, DestType) \
383381
template <typename T, typename R, rounding_mode roundingMode, \
384382
typename OpenCLT, typename OpenCLR> \
385-
detail::enable_if_t< \
386-
is_sint_to_float<T, R>::value && std::is_same<R, DestType>::value, R> \
383+
detail::enable_if_t<is_sint_to_float<T, R>::value && \
384+
(std::is_same<OpenCLR, DestType>::value || \
385+
std::is_same<OpenCLR, _Float16>::value && \
386+
std::is_same<DestType, half>::value), \
387+
R> \
387388
convertImpl(T Value) { \
388389
OpenCLT OpValue = cl::sycl::detail::convertDataToType<T, OpenCLT>(Value); \
389390
return __spirv_Convert##SPIRVOp##_R##DestType(OpValue); \
@@ -399,8 +400,11 @@ __SYCL_GENERATE_CONVERT_IMPL(SToF, double)
399400
#define __SYCL_GENERATE_CONVERT_IMPL(SPIRVOp, DestType) \
400401
template <typename T, typename R, rounding_mode roundingMode, \
401402
typename OpenCLT, typename OpenCLR> \
402-
detail::enable_if_t< \
403-
is_uint_to_float<T, R>::value && std::is_same<R, DestType>::value, R> \
403+
detail::enable_if_t<is_uint_to_float<T, R>::value && \
404+
(std::is_same<OpenCLR, DestType>::value || \
405+
std::is_same<OpenCLR, _Float16>::value && \
406+
std::is_same<DestType, half>::value), \
407+
R> \
404408
convertImpl(T Value) { \
405409
OpenCLT OpValue = cl::sycl::detail::convertDataToType<T, OpenCLT>(Value); \
406410
return __spirv_Convert##SPIRVOp##_R##DestType(OpValue); \
@@ -417,9 +421,11 @@ __SYCL_GENERATE_CONVERT_IMPL(UToF, double)
417421
RoundingModeCondition) \
418422
template <typename T, typename R, rounding_mode roundingMode, \
419423
typename OpenCLT, typename OpenCLR> \
420-
detail::enable_if_t<!std::is_same<T, R>::value && \
421-
is_float_to_float<T, R>::value && \
422-
std::is_same<R, DestType>::value && \
424+
detail::enable_if_t<is_float_to_float<T, R>::value && \
425+
!std::is_same<OpenCLT, OpenCLR>::value && \
426+
(std::is_same<OpenCLR, DestType>::value || \
427+
std::is_same<OpenCLR, _Float16>::value && \
428+
std::is_same<DestType, half>::value) && \
423429
RoundingModeCondition<roundingMode>::value, \
424430
R> \
425431
convertImpl(T Value) { \

‎sycl/test/basic_tests/vec_convert_i_to_i.cpp

Lines changed: 9 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -47,5 +47,14 @@ int main() {
4747
test<ulong, int, 8, rounding_mode::automatic>(
4848
ulong8{3000000000, 2147483647, 100, 150, 2147483648, 1000, 0, 1},
4949
int8{-1294967296, 2147483647, 100, 150, -2147483648, 1000, 0, 1});
50+
51+
test<longlong, ulonglong, 1, rounding_mode::automatic>(
52+
longlong{1},
53+
ulonglong{1});
54+
55+
test<ulonglong, longlong, 1, rounding_mode::automatic>(
56+
ulonglong{1},
57+
longlong{1});
58+
5059
return 0;
5160
}

0 commit comments

Comments
 (0)