Skip to content

Commit b4998f2

Browse files
sndmitrievv-klochkov
sndmitriev
authored and
v-klochkov
committed
[SYCL] Use correct scalar types for vector elements on Windows [revised]
Implementation uses wrong types for 64-bit integers on Windows for vector elements because '[unsigned] long' is being used. But on Windows 'long' has 32 bits as opposed to 64 bits on Linux. This patch corrects this problem. This is a revised fix for the earlier commit e933485. Previous change is reverted. Signed-off-by: Sergey Dmitriev <[email protected]>
1 parent a784071 commit b4998f2

File tree

6 files changed

+101
-34
lines changed

6 files changed

+101
-34
lines changed

‎sycl/include/CL/sycl/detail/generic_type_traits.hpp

Lines changed: 4 additions & 4 deletions
Original file line numberDiff line numberDiff line change
@@ -233,23 +233,23 @@ using is_genint =
233233
// ulongn: ulong2, ulong3, ulong4, ulong8,ulong16
234234
template <typename T>
235235
using is_ulongn = typename is_contained<
236-
T, type_list<cl_ulong2, cl_ulong3, cl_ulong4, cl_ulong8, cl_ulong16>>::type;
236+
T, type_list<ulong2, ulong3, ulong4, ulong8, ulong16>>::type;
237237

238238
// ugenlong: unsigned long int, ulongn
239239
template <typename T>
240240
using is_ugenlong =
241-
std::integral_constant<bool, is_contained<T, type_list<cl_ulong>>::value ||
241+
std::integral_constant<bool, is_contained<T, type_list<ulong>>::value ||
242242
is_ulongn<T>::value>;
243243

244244
// longn: long2, long3, long4, long8, long16
245245
template <typename T>
246246
using is_longn = typename is_contained<
247-
T, type_list<cl_long2, cl_long3, cl_long4, cl_long8, cl_long16>>::type;
247+
T, type_list<long2, long3, long4, long8, long16>>::type;
248248

249249
// genlong: long int, longn
250250
template <typename T>
251251
using is_genlong =
252-
std::integral_constant<bool, is_contained<T, type_list<cl_long>>::value ||
252+
std::integral_constant<bool, is_contained<T, type_list<long>>::value ||
253253
is_longn<T>::value>;
254254

255255
// ulonglongn: ulonglong2, ulonglong3, ulonglong4,ulonglong8, ulonglong16

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

Lines changed: 27 additions & 4 deletions
Original file line numberDiff line numberDiff line change
@@ -1762,7 +1762,7 @@ DECLARE_TYPE_T(half);
17621762
DECLARE_TYPE_T(double);
17631763

17641764
#define GET_CL_TYPE(target, num) __##target##num##_vec_t
1765-
#define GET_SCALAR_CL_TYPE(target) __##target##_t
1765+
#define GET_SCALAR_CL_TYPE(target) target
17661766

17671767
#undef DECLARE_TYPE_T
17681768
#undef DECLARE_LLTYPE_T
@@ -1925,7 +1925,7 @@ using select_apply_cl_t =
19251925
DECLARE_LONGLONG_CONVERTER(base, 16) \
19261926
template <> class BaseCLTypeConverter<base##long, 1> { \
19271927
public: \
1928-
using DataType = GET_SCALAR_CL_TYPE(base); \
1928+
using DataType = base##long; \
19291929
}; \
19301930
} // namespace detail
19311931

@@ -1965,6 +1965,18 @@ using select_apply_cl_t =
19651965
using half3 = vec<half, 3>; \
19661966
using half2 = vec<half, 2>;
19671967

1968+
#define DECLARE_SYCL_VEC_LONG_WO_CONVERTERS(base) \
1969+
using cl_##base##16 = vec<cl_##base, 16>; \
1970+
using cl_##base##8 = vec<cl_##base, 8>; \
1971+
using cl_##base##4 = vec<cl_##base, 4>; \
1972+
using cl_##base##3 = vec<cl_##base, 3>; \
1973+
using cl_##base##2 = vec<cl_##base, 2>; \
1974+
using base##16 = vec<base, 16>; \
1975+
using base##8 = vec<base, 8>; \
1976+
using base##4 = vec<base, 4>; \
1977+
using base##3 = vec<base, 3>; \
1978+
using base##2 = vec<base, 2>;
1979+
19681980
// cl_longlong/cl_ulonglong are not supported in SYCL
19691981
#define DECLARE_SYCL_VEC_LONGLONG_WO_CONVERTERS(base) \
19701982
using base##long16 = vec<base##long, 16>; \
@@ -1989,6 +2001,14 @@ using select_apply_cl_t =
19892001
DECLARE_VECTOR_CONVERTERS(char) \
19902002
DECLARE_SYCL_VEC_CHAR_WO_CONVERTERS
19912003

2004+
#define DECLARE_SYCL_VEC_LONG \
2005+
DECLARE_SIGNED_INTEGRAL_VECTOR_CONVERTERS(long) \
2006+
DECLARE_SYCL_VEC_LONG_WO_CONVERTERS(long)
2007+
2008+
#define DECLARE_SYCL_VEC_ULONG \
2009+
DECLARE_UNSIGNED_INTEGRAL_VECTOR_CONVERTERS(ulong) \
2010+
DECLARE_SYCL_VEC_LONG_WO_CONVERTERS(ulong)
2011+
19922012
#define DECLARE_SYCL_VEC_LONGLONG(base) \
19932013
DECLARE_VECTOR_LONGLONG_CONVERTERS(base) \
19942014
DECLARE_SYCL_VEC_LONGLONG_WO_CONVERTERS(base)
@@ -2004,8 +2024,8 @@ DECLARE_SYCL_SIGNED_INTEGRAL_VEC(short)
20042024
DECLARE_SYCL_UNSIGNED_INTEGRAL_VEC(ushort)
20052025
DECLARE_SYCL_SIGNED_INTEGRAL_VEC(int)
20062026
DECLARE_SYCL_UNSIGNED_INTEGRAL_VEC(uint)
2007-
DECLARE_SYCL_SIGNED_INTEGRAL_VEC(long)
2008-
DECLARE_SYCL_UNSIGNED_INTEGRAL_VEC(ulong)
2027+
DECLARE_SYCL_VEC_LONG
2028+
DECLARE_SYCL_VEC_ULONG
20092029
DECLARE_SYCL_VEC_LONGLONG(long)
20102030
DECLARE_SYCL_VEC_LONGLONG(ulong)
20112031
DECLARE_SYCL_VEC_HALF(half)
@@ -2018,6 +2038,9 @@ DECLARE_SYCL_FLOAT_VEC(double)
20182038
#undef DECLARE_VECTOR_CONVERTERS
20192039
#undef DECLARE_SYCL_VEC
20202040
#undef DECLARE_SYCL_VEC_WO_CONVERTERS
2041+
#undef DECLARE_SYCL_VEC_LONG_WO_CONVERTERS
2042+
#undef DECLARE_SYCL_VEC_LONG
2043+
#undef DECLARE_SYCL_VEC_ULONG
20212044

20222045
} // namespace sycl
20232046
} // namespace cl

‎sycl/test/basic_tests/generic_type_traits.cpp

Lines changed: 40 additions & 4 deletions
Original file line numberDiff line numberDiff line change
@@ -96,12 +96,45 @@ int main() {
9696
is_ugenint
9797
is_intn
9898
is_genint
99+
*/
99100

100-
is_ulongn
101-
is_ugenlong
102-
is_longn
103-
is_genlong
101+
// is_ulongn
102+
static_assert(d::is_ulongn<s::uchar>::value == false, "");
103+
static_assert(d::is_ulongn<s::ulong>::value == false, "");
104+
static_assert(d::is_ulongn<s::ulong2>::value == true, "");
105+
static_assert(d::is_ulongn<s::ulong3>::value == true, "");
106+
static_assert(d::is_ulongn<s::ulong4>::value == true, "");
107+
static_assert(d::is_ulongn<s::ulong8>::value == true, "");
108+
static_assert(d::is_ulongn<s::ulong16>::value == true, "");
109+
110+
// is_ugenlong
111+
static_assert(d::is_ugenlong<s::uchar>::value == false, "");
112+
static_assert(d::is_ugenlong<s::ulong>::value == true, "");
113+
static_assert(d::is_ugenlong<s::ulong2>::value == true, "");
114+
static_assert(d::is_ugenlong<s::ulong3>::value == true, "");
115+
static_assert(d::is_ugenlong<s::ulong4>::value == true, "");
116+
static_assert(d::is_ugenlong<s::ulong8>::value == true, "");
117+
static_assert(d::is_ugenlong<s::ulong16>::value == true, "");
118+
119+
// is_longn
120+
static_assert(d::is_longn<char>::value == false, "");
121+
static_assert(d::is_longn<long>::value == false, "");
122+
static_assert(d::is_longn<s::long2>::value == true, "");
123+
static_assert(d::is_longn<s::long3>::value == true, "");
124+
static_assert(d::is_longn<s::long4>::value == true, "");
125+
static_assert(d::is_longn<s::long8>::value == true, "");
126+
static_assert(d::is_longn<s::long16>::value == true, "");
127+
128+
// is_genlong
129+
static_assert(d::is_genlong<char>::value == false, "");
130+
static_assert(d::is_genlong<long>::value == true, "");
131+
static_assert(d::is_genlong<s::long2>::value == true, "");
132+
static_assert(d::is_genlong<s::long3>::value == true, "");
133+
static_assert(d::is_genlong<s::long4>::value == true, "");
134+
static_assert(d::is_genlong<s::long8>::value == true, "");
135+
static_assert(d::is_genlong<s::long16>::value == true, "");
104136

137+
/*
105138
is_ulonglongn
106139
is_ugenlonglong
107140
is_longlongn
@@ -143,6 +176,9 @@ int main() {
143176
static_assert(d::is_nan_type<unsigned long long int>::value == true, "");
144177
static_assert(d::is_nan_type<s::longlong>::value == false, "");
145178
static_assert(d::is_nan_type<s::ulonglong>::value == true, "");
179+
static_assert(d::is_nan_type<unsigned long>::value == true, "");
180+
static_assert(d::is_nan_type<long>::value == false, "");
181+
static_assert(d::is_nan_type<s::ulong>::value == true, "");
146182
/*
147183
float_point_to_sign_integeral
148184

‎sycl/test/basic_tests/vectors/vector_ulonglong.cpp

Lines changed: 0 additions & 17 deletions
This file was deleted.

‎sycl/test/basic_tests/vectors/vectors.cpp

Lines changed: 20 additions & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -1,4 +1,4 @@
1-
// RUN: %clangxx -fsycl %s -o %t.out -lOpenCL
1+
// RUN: %clangxx %s -o %t.out -lOpenCL -lsycl
22
// RUN: %t.out
33
//==--------------- vectors.cpp - SYCL vectors test ------------------------==//
44
//
@@ -57,5 +57,24 @@ int main() {
5757
assert(static_cast<float>(b_vec.z()) == static_cast<float>(0.5));
5858
assert(static_cast<float>(b_vec.w()) == static_cast<float>(0.5));
5959

60+
// Check that vector with 'unsigned long long' elements has enough bits to
61+
// store value.
62+
unsigned long long ull_ref = 1ull - 2ull;
63+
auto ull_vec = cl::sycl::vec<unsigned long long, 1>(ull_ref);
64+
unsigned long long ull_val = ull_vec.template swizzle<cl::sycl::elem::s0>();
65+
assert(ull_val == ull_ref);
66+
67+
// Check that [u]long[n] type aliases match vec<[unsigned] long, n> types.
68+
assert((std::is_same<cl::sycl::vec<long, 2>, cl::sycl::long2>::value));
69+
assert((std::is_same<cl::sycl::vec<long, 3>, cl::sycl::long3>::value));
70+
assert((std::is_same<cl::sycl::vec<long, 4>, cl::sycl::long4>::value));
71+
assert((std::is_same<cl::sycl::vec<long, 8>, cl::sycl::long8>::value));
72+
assert((std::is_same<cl::sycl::vec<long, 16>, cl::sycl::long16>::value));
73+
assert((std::is_same<cl::sycl::vec<unsigned long, 2>, cl::sycl::ulong2>::value));
74+
assert((std::is_same<cl::sycl::vec<unsigned long, 3>, cl::sycl::ulong3>::value));
75+
assert((std::is_same<cl::sycl::vec<unsigned long, 4>, cl::sycl::ulong4>::value));
76+
assert((std::is_same<cl::sycl::vec<unsigned long, 8>, cl::sycl::ulong8>::value));
77+
assert((std::is_same<cl::sycl::vec<unsigned long, 16>, cl::sycl::ulong16>::value));
78+
6079
return 0;
6180
}

‎sycl/test/built-ins/vector_math.cpp

Lines changed: 10 additions & 4 deletions
Original file line numberDiff line numberDiff line change
@@ -206,18 +206,24 @@ int main() {
206206

207207
// nan (ulong)
208208
{
209-
s::cl_double2 r{ 0 };
209+
// Result type depends on the ulong type size which varies across targets.
210+
// It is a 64-bit type on Linux and 32-bit on Windows.
211+
using res_scalar_t = std::conditional<sizeof(s::ulong) == sizeof(float),
212+
s::cl_float, s::cl_double>::type;
213+
using res_vector_t = s::vec<res_scalar_t, 2>;
214+
215+
res_vector_t r{ 0 };
210216
{
211-
s::buffer<s::cl_double2, 1> BufR(&r, s::range<1>(1));
217+
s::buffer<res_vector_t, 1> BufR(&r, s::range<1>(1));
212218
s::queue myQueue;
213219
myQueue.submit([&](s::handler &cgh) {
214220
auto AccR = BufR.get_access<s::access::mode::write>(cgh);
215221
s::ulong2 in{1, 1};
216222
cgh.single_task<class nanISUL2>([=]() { AccR[0] = s::nan(in); });
217223
});
218224
}
219-
s::cl_double x = r.x();
220-
s::cl_double y = r.y();
225+
res_scalar_t x = r.x();
226+
res_scalar_t y = r.y();
221227
assert(std::isnan(x));
222228
assert(std::isnan(y));
223229
}

0 commit comments

Comments
 (0)