Skip to content

Commit 62db4e5

Browse files
committed
Convert to long long correction
Signed-off-by: Aleksander Fadeev <[email protected]>
1 parent 6d204e0 commit 62db4e5

File tree

2 files changed

+65
-4
lines changed

2 files changed

+65
-4
lines changed

sycl/include/CL/sycl/types.hpp

Lines changed: 7 additions & 3 deletions
Original file line numberDiff line numberDiff line change
@@ -323,8 +323,10 @@ using Rtn = detail::bool_constant<Mode == rounding_mode::rtn>;
323323
template <typename T, typename R, rounding_mode roundingMode, typename OpenCLT, typename OpenCLR> \
324324
detail::enable_if_t< \
325325
!std::is_same<T, R>::value && is_sint_to_sint<T, R>::value && \
326-
std::is_same<OpenCLR, DestType>::value && \
327-
!std::is_same<OpenCLT, OpenCLR>::value, \
326+
(std::is_same<OpenCLR, DestType>::value || \
327+
std::is_same<OpenCLR, signed char>::value && \
328+
std::is_same<DestType, char>::value) && \
329+
!std::is_same<OpenCLT, OpenCLR>::value, \
328330
R> \
329331
convertImpl(T Value) { \
330332
OpenCLT OpValue = cl::sycl::detail::convertDataToType<T, OpenCLT>(Value); \
@@ -429,7 +431,9 @@ __SYCL_GENERATE_CONVERT_IMPL_FOR_ROUNDING_MODE(rtn, Rtn)
429431
RoundingModeCondition) \
430432
template <typename T, typename R, rounding_mode roundingMode, typename OpenCLT, typename OpenCLR> \
431433
detail::enable_if_t<is_float_to_int<T, R>::value && \
432-
std::is_same<OpenCLR, DestType>::value && \
434+
(std::is_same<OpenCLR, DestType>::value || \
435+
std::is_same<OpenCLR, signed char>::value && \
436+
std::is_same<DestType, char>::value) && \
433437
RoundingModeCondition<roundingMode>::value, \
434438
R> \
435439
convertImpl(T Value) { \

sycl/test/basic_tests/vec_convert_half.cpp

Lines changed: 58 additions & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -10,10 +10,67 @@
1010
//
1111
//===----------------------------------------------------------------------===//
1212

13-
#include "vec_convert.hpp"
13+
#include <CL/sycl.hpp>
14+
15+
#include <cassert>
1416

1517
// TODO make the convertion on CPU and HOST identical
1618

19+
using namespace cl::sycl;
20+
21+
template <typename T, typename convertT, int roundingMode>
22+
class kernel_name;
23+
24+
template <int N>
25+
struct helper;
26+
27+
template <>
28+
struct helper<0> {
29+
template <typename T, int NumElements>
30+
static void compare(const vec<T, NumElements> &x,
31+
const vec<T, NumElements> &y) {
32+
const T xs = x.template swizzle<0>();
33+
const T ys = y.template swizzle<0>();
34+
assert(xs == ys);
35+
}
36+
};
37+
38+
template <int N>
39+
struct helper {
40+
template <typename T, int NumElements>
41+
static void compare(const vec<T, NumElements> &x,
42+
const vec<T, NumElements> &y) {
43+
const T xs = x.template swizzle<N>();
44+
const T ys = y.template swizzle<N>();
45+
helper<N - 1>::compare(x, y);
46+
assert(xs == ys);
47+
}
48+
};
49+
50+
template <typename T, typename convertT, int NumElements,
51+
rounding_mode roundingMode>
52+
void test(const vec<T, NumElements> &ToConvert,
53+
const vec<convertT, NumElements> &Expected) {
54+
vec<convertT, NumElements> Converted{0};
55+
{
56+
buffer<vec<convertT, NumElements>, 1> Buffer{&Converted, range<1>{1}};
57+
queue Queue;
58+
59+
cl::sycl::device D = Queue.get_device();
60+
if (!D.has_extension("cl_khr_fp16"))
61+
exit(0);
62+
63+
Queue.submit([&](handler &CGH) {
64+
accessor<vec<convertT, NumElements>, 1, access::mode::write> Accessor(
65+
Buffer, CGH);
66+
CGH.single_task<class kernel_name<T, convertT, static_cast<int>(roundingMode)>>([=]() {
67+
Accessor[0] = ToConvert.template convert<convertT, roundingMode>();
68+
});
69+
});
70+
}
71+
helper<NumElements - 1>::compare(Converted, Expected);
72+
}
73+
1774
int main() {
1875
//automatic
1976
test<double, half, 4, rounding_mode::automatic>(

0 commit comments

Comments
 (0)