Skip to content

Commit 11d0d87

Browse files
cperkinsintelcperkinsintel
committed
Revert "[SYCL] Fix sycl::vec unary ops (intel#10722)"
This reverts commit f7b00b7.
1 parent 292309c commit 11d0d87

File tree

2 files changed

+49
-154
lines changed

2 files changed

+49
-154
lines changed

‎sycl/include/sycl/types.hpp

Lines changed: 49 additions & 91 deletions
Original file line numberDiff line numberDiff line change
@@ -580,17 +580,13 @@ template <typename Type, int NumElements> class vec {
580580
// vector extension. This is for MSVC compatibility, which has a max alignment
581581
// of 64 for direct params. If we drop MSVC, we can have alignment the same as
582582
// size and use vector extensions for all sizes.
583-
static constexpr bool IsUsingArrayOnDevice =
583+
static constexpr bool IsUsingArray =
584584
(IsHostHalf || IsSizeGreaterThanMaxAlign);
585585

586586
#if defined(__SYCL_DEVICE_ONLY__)
587-
static constexpr bool NativeVec = NumElements > 1 && !IsUsingArrayOnDevice;
588-
static constexpr bool IsUsingArrayOnHost =
589-
false; // we are not compiling for host.
587+
static constexpr bool NativeVec = NumElements > 1 && !IsUsingArray;
590588
#else
591589
static constexpr bool NativeVec = false;
592-
static constexpr bool IsUsingArrayOnHost =
593-
true; // host always uses std::array.
594590
#endif
595591

596592
static constexpr int getNumElements() { return NumElements; }
@@ -771,15 +767,6 @@ template <typename Type, int NumElements> class vec {
771767
return *this;
772768
}
773769

774-
template <typename T = void>
775-
using EnableIfUsingArray =
776-
typename std::enable_if_t<IsUsingArrayOnDevice || IsUsingArrayOnHost, T>;
777-
778-
template <typename T = void>
779-
using EnableIfNotUsingArray =
780-
typename std::enable_if_t<!IsUsingArrayOnDevice && !IsUsingArrayOnHost,
781-
T>;
782-
783770
#ifdef __SYCL_DEVICE_ONLY__
784771
template <typename T = void>
785772
using EnableIfNotHostHalf = typename std::enable_if_t<!IsHostHalf, T>;
@@ -788,29 +775,27 @@ template <typename Type, int NumElements> class vec {
788775
using EnableIfHostHalf = typename std::enable_if_t<IsHostHalf, T>;
789776

790777
template <typename T = void>
791-
using EnableIfUsingArrayOnDevice =
792-
typename std::enable_if_t<IsUsingArrayOnDevice, T>;
778+
using EnableIfUsingArray = typename std::enable_if_t<IsUsingArray, T>;
793779

794780
template <typename T = void>
795-
using EnableIfNotUsingArrayOnDevice =
796-
typename std::enable_if_t<!IsUsingArrayOnDevice, T>;
781+
using EnableIfNotUsingArray = typename std::enable_if_t<!IsUsingArray, T>;
797782

798783
template <typename Ty = DataT>
799-
explicit constexpr vec(const EnableIfNotUsingArrayOnDevice<Ty> &arg)
784+
explicit constexpr vec(const EnableIfNotUsingArray<Ty> &arg)
800785
: m_Data{DataType(vec_data<Ty>::get(arg))} {}
801786

802787
template <typename Ty = DataT>
803788
typename std::enable_if_t<
804789
std::is_fundamental_v<vec_data_t<Ty>> ||
805790
std::is_same_v<typename std::remove_const_t<Ty>, half>,
806791
vec &>
807-
operator=(const EnableIfNotUsingArrayOnDevice<Ty> &Rhs) {
792+
operator=(const EnableIfNotUsingArray<Ty> &Rhs) {
808793
m_Data = (DataType)vec_data<Ty>::get(Rhs);
809794
return *this;
810795
}
811796

812797
template <typename Ty = DataT>
813-
explicit constexpr vec(const EnableIfUsingArrayOnDevice<Ty> &arg)
798+
explicit constexpr vec(const EnableIfUsingArray<Ty> &arg)
814799
: vec{detail::RepeatValue<NumElements>(
815800
static_cast<vec_data_t<DataT>>(arg)),
816801
std::make_index_sequence<NumElements>()} {}
@@ -820,7 +805,7 @@ template <typename Type, int NumElements> class vec {
820805
std::is_fundamental_v<vec_data_t<Ty>> ||
821806
std::is_same_v<typename std::remove_const_t<Ty>, half>,
822807
vec &>
823-
operator=(const EnableIfUsingArrayOnDevice<Ty> &Rhs) {
808+
operator=(const EnableIfUsingArray<Ty> &Rhs) {
824809
for (int i = 0; i < NumElements; ++i) {
825810
setValue(i, Rhs);
826811
}
@@ -856,22 +841,22 @@ template <typename Type, int NumElements> class vec {
856841
std::is_convertible_v<T, DataT> && NumElements == IdxNum, DataT>;
857842
template <typename Ty = DataT>
858843
constexpr vec(const EnableIfMultipleElems<2, Ty> Arg0,
859-
const EnableIfNotUsingArrayOnDevice<Ty> Arg1)
844+
const EnableIfNotUsingArray<Ty> Arg1)
860845
: m_Data{vec_data<Ty>::get(Arg0), vec_data<Ty>::get(Arg1)} {}
861846
template <typename Ty = DataT>
862847
constexpr vec(const EnableIfMultipleElems<3, Ty> Arg0,
863-
const EnableIfNotUsingArrayOnDevice<Ty> Arg1, const DataT Arg2)
848+
const EnableIfNotUsingArray<Ty> Arg1, const DataT Arg2)
864849
: m_Data{vec_data<Ty>::get(Arg0), vec_data<Ty>::get(Arg1),
865850
vec_data<Ty>::get(Arg2)} {}
866851
template <typename Ty = DataT>
867852
constexpr vec(const EnableIfMultipleElems<4, Ty> Arg0,
868-
const EnableIfNotUsingArrayOnDevice<Ty> Arg1, const DataT Arg2,
853+
const EnableIfNotUsingArray<Ty> Arg1, const DataT Arg2,
869854
const Ty Arg3)
870855
: m_Data{vec_data<Ty>::get(Arg0), vec_data<Ty>::get(Arg1),
871856
vec_data<Ty>::get(Arg2), vec_data<Ty>::get(Arg3)} {}
872857
template <typename Ty = DataT>
873858
constexpr vec(const EnableIfMultipleElems<8, Ty> Arg0,
874-
const EnableIfNotUsingArrayOnDevice<Ty> Arg1, const DataT Arg2,
859+
const EnableIfNotUsingArray<Ty> Arg1, const DataT Arg2,
875860
const DataT Arg3, const DataT Arg4, const DataT Arg5,
876861
const DataT Arg6, const DataT Arg7)
877862
: m_Data{vec_data<Ty>::get(Arg0), vec_data<Ty>::get(Arg1),
@@ -880,7 +865,7 @@ template <typename Type, int NumElements> class vec {
880865
vec_data<Ty>::get(Arg6), vec_data<Ty>::get(Arg7)} {}
881866
template <typename Ty = DataT>
882867
constexpr vec(const EnableIfMultipleElems<16, Ty> Arg0,
883-
const EnableIfNotUsingArrayOnDevice<Ty> Arg1, const DataT Arg2,
868+
const EnableIfNotUsingArray<Ty> Arg1, const DataT Arg2,
884869
const DataT Arg3, const DataT Arg4, const DataT Arg5,
885870
const DataT Arg6, const DataT Arg7, const DataT Arg8,
886871
const DataT Arg9, const DataT ArgA, const DataT ArgB,
@@ -910,15 +895,15 @@ template <typename Type, int NumElements> class vec {
910895
typename std::enable_if_t<std::is_same_v<vector_t_, vector_t> &&
911896
!std::is_same_v<vector_t_, DataT>>>
912897
constexpr vec(vector_t openclVector) {
913-
if constexpr (!IsUsingArrayOnDevice) {
898+
if constexpr (!IsUsingArray) {
914899
m_Data = openclVector;
915900
} else {
916901
m_Data = bit_cast<DataType>(openclVector);
917902
}
918903
}
919904

920905
operator vector_t() const {
921-
if constexpr (!IsUsingArrayOnDevice) {
906+
if constexpr (!IsUsingArray) {
922907
return m_Data;
923908
} else {
924909
auto ptr = bit_cast<const VectorDataType *>((&m_Data)->data());
@@ -1079,7 +1064,7 @@ template <typename Type, int NumElements> class vec {
10791064
#ifdef __SYCL_DEVICE_ONLY__
10801065
#define __SYCL_BINOP(BINOP, OPASSIGN, CONVERT) \
10811066
template <typename Ty = vec> \
1082-
vec operator BINOP(const EnableIfNotUsingArrayOnDevice<Ty> &Rhs) const { \
1067+
vec operator BINOP(const EnableIfNotUsingArray<Ty> &Rhs) const { \
10831068
vec Ret; \
10841069
Ret.m_Data = m_Data BINOP Rhs.m_Data; \
10851070
if constexpr (std::is_same_v<Type, bool> && CONVERT) { \
@@ -1088,7 +1073,7 @@ template <typename Type, int NumElements> class vec {
10881073
return Ret; \
10891074
} \
10901075
template <typename Ty = vec> \
1091-
vec operator BINOP(const EnableIfUsingArrayOnDevice<Ty> &Rhs) const { \
1076+
vec operator BINOP(const EnableIfUsingArray<Ty> &Rhs) const { \
10921077
vec Ret; \
10931078
for (size_t I = 0; I < NumElements; ++I) { \
10941079
Ret.setValue(I, (getValue(I) BINOP Rhs.getValue(I))); \
@@ -1242,94 +1227,67 @@ template <typename Type, int NumElements> class vec {
12421227
__SYCL_UOP(--, -=)
12431228
#undef __SYCL_UOP
12441229

1245-
// operator~() available only when: dataT != float && dataT != double
1246-
// && dataT != half
1230+
// Available only when: dataT != cl_float && dataT != cl_double
1231+
// && dataT != cl_half
12471232
template <typename T = DataT>
1248-
typename std::enable_if_t<!std::is_floating_point_v<vec_data_t<T>> &&
1249-
(!IsUsingArrayOnDevice && !IsUsingArrayOnHost),
1250-
vec>
1233+
typename std::enable_if_t<std::is_integral_v<vec_data_t<T>>, vec>
12511234
operator~() const {
1235+
// Use __SYCL_DEVICE_ONLY__ macro because cast to OpenCL vector type is defined
1236+
// by SYCL device compiler only.
1237+
#ifdef __SYCL_DEVICE_ONLY__
12521238
vec Ret{(typename vec::DataType) ~m_Data};
12531239
if constexpr (std::is_same_v<Type, bool>) {
12541240
Ret.ConvertToDataT();
12551241
}
12561242
return Ret;
1257-
}
1258-
template <typename T = DataT>
1259-
typename std::enable_if_t<!std::is_floating_point_v<vec_data_t<T>> &&
1260-
(IsUsingArrayOnDevice || IsUsingArrayOnHost),
1261-
vec>
1262-
operator~() const {
1243+
#else
12631244
vec Ret{};
12641245
for (size_t I = 0; I < NumElements; ++I) {
12651246
Ret.setValue(I, ~getValue(I));
12661247
}
12671248
return Ret;
1249+
#endif
12681250
}
12691251

1270-
// operator!
1271-
template <typename T = DataT, int N = NumElements>
1272-
EnableIfNotUsingArray<vec<T, N>> operator!() const {
1273-
return vec<T, N>{(typename vec<DataT, NumElements>::DataType) !m_Data};
1274-
}
1275-
1276-
// std::byte neither supports ! unary op or casting, so special handling is
1277-
// needed. And, worse, Windows has a conflict with 'byte'.
1278-
#if (!defined(_HAS_STD_BYTE) || _HAS_STD_BYTE != 0)
1279-
template <typename T = DataT, int N = NumElements>
1280-
typename std::enable_if_t<std::is_same_v<std::byte, T> &&
1281-
(IsUsingArrayOnDevice || IsUsingArrayOnHost),
1282-
vec<T, N>>
1283-
operator!() const {
1284-
vec Ret{};
1285-
for (size_t I = 0; I < NumElements; ++I) {
1286-
Ret.setValue(I, std::byte{!vec_data<DataT>::get(getValue(I))});
1287-
}
1288-
return Ret;
1289-
}
1290-
1291-
template <typename T = DataT, int N = NumElements>
1292-
typename std::enable_if_t<!std::is_same_v<std::byte, T> &&
1293-
(IsUsingArrayOnDevice || IsUsingArrayOnHost),
1294-
vec<T, N>>
1295-
operator!() const {
1296-
vec Ret{};
1297-
for (size_t I = 0; I < NumElements; ++I)
1298-
Ret.setValue(I, !vec_data<DataT>::get(getValue(I)));
1299-
return Ret;
1300-
}
1252+
vec<rel_t, NumElements> operator!() const {
1253+
// Use __SYCL_DEVICE_ONLY__ macro because cast to OpenCL vector type is defined
1254+
// by SYCL device compiler only.
1255+
#ifdef __SYCL_DEVICE_ONLY__
1256+
return vec<rel_t, NumElements>{
1257+
(typename vec<rel_t, NumElements>::DataType) !m_Data};
13011258
#else
1302-
template <typename T = DataT, int N = NumElements>
1303-
EnableIfUsingArray<vec<T, N>> operator!() const {
1304-
vec Ret{};
1305-
for (size_t I = 0; I < NumElements; ++I)
1259+
vec<rel_t, NumElements> Ret{};
1260+
for (size_t I = 0; I < NumElements; ++I) {
13061261
Ret.setValue(I, !vec_data<DataT>::get(getValue(I)));
1262+
}
13071263
return Ret;
1308-
}
13091264
#endif
1310-
1311-
// operator +
1312-
template <typename T = vec> EnableIfNotUsingArray<T> operator+() const {
1313-
return vec{+m_Data};
13141265
}
13151266

1316-
template <typename T = vec> EnableIfUsingArray<T> operator+() const {
1267+
vec operator+() const {
1268+
// Use __SYCL_DEVICE_ONLY__ macro because cast to OpenCL vector type is defined
1269+
// by SYCL device compiler only.
1270+
#ifdef __SYCL_DEVICE_ONLY__
1271+
return vec{+m_Data};
1272+
#else
13171273
vec Ret{};
13181274
for (size_t I = 0; I < NumElements; ++I)
13191275
Ret.setValue(I, vec_data<DataT>::get(+vec_data<DataT>::get(getValue(I))));
13201276
return Ret;
1277+
#endif
13211278
}
13221279

1323-
// operator -
1324-
template <typename T = vec> EnableIfNotUsingArray<T> operator-() const {
1280+
vec operator-() const {
1281+
// Use __SYCL_DEVICE_ONLY__ macro because cast to OpenCL vector type is defined
1282+
// by SYCL device compiler only.
1283+
#ifdef __SYCL_DEVICE_ONLY__
13251284
return vec{-m_Data};
1326-
}
1327-
1328-
template <typename T = vec> EnableIfUsingArray<T> operator-() const {
1285+
#else
13291286
vec Ret{};
13301287
for (size_t I = 0; I < NumElements; ++I)
13311288
Ret.setValue(I, vec_data<DataT>::get(-vec_data<DataT>::get(getValue(I))));
13321289
return Ret;
1290+
#endif
13331291
}
13341292

13351293
// OP is: &&, ||
@@ -1345,7 +1303,7 @@ template <typename Type, int NumElements> class vec {
13451303
template <template <typename> class Operation,
13461304
typename Ty = vec<DataT, NumElements>>
13471305
vec<DataT, NumElements>
1348-
operatorHelper(const EnableIfNotUsingArrayOnDevice<Ty> &Rhs) const {
1306+
operatorHelper(const EnableIfNotUsingArray<Ty> &Rhs) const {
13491307
vec<DataT, NumElements> Result;
13501308
Operation<DataType> Op;
13511309
Result.m_Data = Op(m_Data, Rhs.m_Data);
@@ -1355,7 +1313,7 @@ template <typename Type, int NumElements> class vec {
13551313
template <template <typename> class Operation,
13561314
typename Ty = vec<DataT, NumElements>>
13571315
vec<DataT, NumElements>
1358-
operatorHelper(const EnableIfUsingArrayOnDevice<Ty> &Rhs) const {
1316+
operatorHelper(const EnableIfUsingArray<Ty> &Rhs) const {
13591317
vec<DataT, NumElements> Result;
13601318
Operation<DataT> Op;
13611319
for (size_t I = 0; I < NumElements; ++I) {

‎sycl/test/basic_tests/types.cpp

Lines changed: 0 additions & 63 deletions
Original file line numberDiff line numberDiff line change
@@ -101,67 +101,6 @@ template <> inline void checkSizeForFloatingPoint<s::half, sizeof(int16_t)>() {
101101
static_assert(sizeof(s::half) == sizeof(int16_t), "");
102102
}
103103

104-
template <typename vecType, int numOfElems>
105-
std::string vec2string(const sycl::vec<vecType, numOfElems> &vec) {
106-
std::string str = "";
107-
for (size_t i = 0; i < numOfElems - 1; ++i) {
108-
str += std::to_string(vec[i]) + ",";
109-
}
110-
str = "{" + str + std::to_string(vec[numOfElems - 1]) + "}";
111-
return str;
112-
}
113-
114-
// the math built-in testing ensures that the vec binary ops get tested,
115-
// but the unary ops are only tested by the CTS tests. Here we do some
116-
// basic testing of the unary ops, ensuring they compile correctly.
117-
template <typename T> void checkVecUnaryOps(T &v) {
118-
119-
std::cout << vec2string(v) << std::endl;
120-
121-
T d = +v;
122-
std::cout << vec2string(d) << std::endl;
123-
124-
T e = -v;
125-
std::cout << vec2string(e) << std::endl;
126-
127-
// ~ only supported by integral types.
128-
if constexpr (std::is_integral_v<T>) {
129-
T g = ~v;
130-
std::cout << vec2string(g) << std::endl;
131-
}
132-
133-
T f = !v;
134-
std::cout << vec2string(f) << std::endl;
135-
}
136-
137-
void checkVariousVecUnaryOps() {
138-
sycl::vec<int, 1> vi1{1};
139-
checkVecUnaryOps(vi1);
140-
sycl::vec<int, 16> vi{1, 2, 0, -4, 1, 2, 0, -4, 1, 2, 0, -4, 1, 2, 0, -4};
141-
checkVecUnaryOps(vi);
142-
143-
sycl::vec<long, 1> vl1{1};
144-
checkVecUnaryOps(vl1);
145-
sycl::vec<long, 16> vl{2, 3, 0, -5, 2, 3, 0, -5, 2, 3, 0, -5, 2, 3, 0, -5};
146-
checkVecUnaryOps(vl);
147-
148-
sycl::vec<long long, 1> vll1{1};
149-
checkVecUnaryOps(vll1);
150-
sycl::vec<long long, 16> vll{0, 3, 4, -6, 0, 3, 4, -6,
151-
0, 3, 4, -6, 0, 3, 4, -6};
152-
checkVecUnaryOps(vll);
153-
154-
sycl::vec<float, 1> vf1{1};
155-
checkVecUnaryOps(vf1);
156-
sycl::vec<float, 16> vf{0, 4, 5, -9, 0, 4, 5, -9, 0, 4, 5, -9, 0, 4, 5, -9};
157-
checkVecUnaryOps(vf);
158-
159-
sycl::vec<double, 1> vd1{1};
160-
checkVecUnaryOps(vd1);
161-
sycl::vec<double, 16> vd{0, 4, 5, -9, 0, 4, 5, -9, 0, 4, 5, -9, 0, 4, 5, -9};
162-
checkVecUnaryOps(vd);
163-
}
164-
165104
int main() {
166105
// Test for creating constexpr expressions
167106
constexpr sycl::specialization_id<sycl::vec<sycl::half, 2>> id(1.0);
@@ -187,7 +126,5 @@ int main() {
187126
checkSizeForFloatingPoint<s::opencl::cl_float, sizeof(int32_t)>();
188127
checkSizeForFloatingPoint<s::opencl::cl_double, sizeof(int64_t)>();
189128

190-
checkVariousVecUnaryOps();
191-
192129
return 0;
193130
}

0 commit comments

Comments
 (0)