Skip to content

Commit 5470741

Browse files
cperkinsintelcperkinsintel
committed
Revert "[SYCL] Fix sycl::vec unary ops (intel#10722)"
This reverts commit f7b00b7.
1 parent 8ac08f1 commit 5470741

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
@@ -582,17 +582,13 @@ template <typename Type, int NumElements> class vec {
582582
// vector extension. This is for MSVC compatibility, which has a max alignment
583583
// of 64 for direct params. If we drop MSVC, we can have alignment the same as
584584
// size and use vector extensions for all sizes.
585-
static constexpr bool IsUsingArrayOnDevice =
585+
static constexpr bool IsUsingArray =
586586
(IsHostHalf || IsSizeGreaterThanMaxAlign);
587587

588588
#if defined(__SYCL_DEVICE_ONLY__)
589-
static constexpr bool NativeVec = NumElements > 1 && !IsUsingArrayOnDevice;
590-
static constexpr bool IsUsingArrayOnHost =
591-
false; // we are not compiling for host.
589+
static constexpr bool NativeVec = NumElements > 1 && !IsUsingArray;
592590
#else
593591
static constexpr bool NativeVec = false;
594-
static constexpr bool IsUsingArrayOnHost =
595-
true; // host always uses std::array.
596592
#endif
597593

598594
static constexpr int getNumElements() { return NumElements; }
@@ -773,15 +769,6 @@ template <typename Type, int NumElements> class vec {
773769
return *this;
774770
}
775771

776-
template <typename T = void>
777-
using EnableIfUsingArray =
778-
typename std::enable_if_t<IsUsingArrayOnDevice || IsUsingArrayOnHost, T>;
779-
780-
template <typename T = void>
781-
using EnableIfNotUsingArray =
782-
typename std::enable_if_t<!IsUsingArrayOnDevice && !IsUsingArrayOnHost,
783-
T>;
784-
785772
#ifdef __SYCL_DEVICE_ONLY__
786773
template <typename T = void>
787774
using EnableIfNotHostHalf = typename std::enable_if_t<!IsHostHalf, T>;
@@ -790,29 +777,27 @@ template <typename Type, int NumElements> class vec {
790777
using EnableIfHostHalf = typename std::enable_if_t<IsHostHalf, T>;
791778

792779
template <typename T = void>
793-
using EnableIfUsingArrayOnDevice =
794-
typename std::enable_if_t<IsUsingArrayOnDevice, T>;
780+
using EnableIfUsingArray = typename std::enable_if_t<IsUsingArray, T>;
795781

796782
template <typename T = void>
797-
using EnableIfNotUsingArrayOnDevice =
798-
typename std::enable_if_t<!IsUsingArrayOnDevice, T>;
783+
using EnableIfNotUsingArray = typename std::enable_if_t<!IsUsingArray, T>;
799784

800785
template <typename Ty = DataT>
801-
explicit constexpr vec(const EnableIfNotUsingArrayOnDevice<Ty> &arg)
786+
explicit constexpr vec(const EnableIfNotUsingArray<Ty> &arg)
802787
: m_Data{DataType(vec_data<Ty>::get(arg))} {}
803788

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

814799
template <typename Ty = DataT>
815-
explicit constexpr vec(const EnableIfUsingArrayOnDevice<Ty> &arg)
800+
explicit constexpr vec(const EnableIfUsingArray<Ty> &arg)
816801
: vec{detail::RepeatValue<NumElements>(
817802
static_cast<vec_data_t<DataT>>(arg)),
818803
std::make_index_sequence<NumElements>()} {}
@@ -822,7 +807,7 @@ template <typename Type, int NumElements> class vec {
822807
std::is_fundamental_v<vec_data_t<Ty>> ||
823808
std::is_same_v<typename std::remove_const_t<Ty>, half>,
824809
vec &>
825-
operator=(const EnableIfUsingArrayOnDevice<Ty> &Rhs) {
810+
operator=(const EnableIfUsingArray<Ty> &Rhs) {
826811
for (int i = 0; i < NumElements; ++i) {
827812
setValue(i, Rhs);
828813
}
@@ -858,22 +843,22 @@ template <typename Type, int NumElements> class vec {
858843
std::is_convertible_v<T, DataT> && NumElements == IdxNum, DataT>;
859844
template <typename Ty = DataT>
860845
constexpr vec(const EnableIfMultipleElems<2, Ty> Arg0,
861-
const EnableIfNotUsingArrayOnDevice<Ty> Arg1)
846+
const EnableIfNotUsingArray<Ty> Arg1)
862847
: m_Data{vec_data<Ty>::get(Arg0), vec_data<Ty>::get(Arg1)} {}
863848
template <typename Ty = DataT>
864849
constexpr vec(const EnableIfMultipleElems<3, Ty> Arg0,
865-
const EnableIfNotUsingArrayOnDevice<Ty> Arg1, const DataT Arg2)
850+
const EnableIfNotUsingArray<Ty> Arg1, const DataT Arg2)
866851
: m_Data{vec_data<Ty>::get(Arg0), vec_data<Ty>::get(Arg1),
867852
vec_data<Ty>::get(Arg2)} {}
868853
template <typename Ty = DataT>
869854
constexpr vec(const EnableIfMultipleElems<4, Ty> Arg0,
870-
const EnableIfNotUsingArrayOnDevice<Ty> Arg1, const DataT Arg2,
855+
const EnableIfNotUsingArray<Ty> Arg1, const DataT Arg2,
871856
const Ty Arg3)
872857
: m_Data{vec_data<Ty>::get(Arg0), vec_data<Ty>::get(Arg1),
873858
vec_data<Ty>::get(Arg2), vec_data<Ty>::get(Arg3)} {}
874859
template <typename Ty = DataT>
875860
constexpr vec(const EnableIfMultipleElems<8, Ty> Arg0,
876-
const EnableIfNotUsingArrayOnDevice<Ty> Arg1, const DataT Arg2,
861+
const EnableIfNotUsingArray<Ty> Arg1, const DataT Arg2,
877862
const DataT Arg3, const DataT Arg4, const DataT Arg5,
878863
const DataT Arg6, const DataT Arg7)
879864
: m_Data{vec_data<Ty>::get(Arg0), vec_data<Ty>::get(Arg1),
@@ -882,7 +867,7 @@ template <typename Type, int NumElements> class vec {
882867
vec_data<Ty>::get(Arg6), vec_data<Ty>::get(Arg7)} {}
883868
template <typename Ty = DataT>
884869
constexpr vec(const EnableIfMultipleElems<16, Ty> Arg0,
885-
const EnableIfNotUsingArrayOnDevice<Ty> Arg1, const DataT Arg2,
870+
const EnableIfNotUsingArray<Ty> Arg1, const DataT Arg2,
886871
const DataT Arg3, const DataT Arg4, const DataT Arg5,
887872
const DataT Arg6, const DataT Arg7, const DataT Arg8,
888873
const DataT Arg9, const DataT ArgA, const DataT ArgB,
@@ -912,15 +897,15 @@ template <typename Type, int NumElements> class vec {
912897
std::is_same<vector_t_, vector_t>::value &&
913898
!std::is_same<vector_t_, DataT>::value>>
914899
constexpr vec(vector_t openclVector) {
915-
if constexpr (!IsUsingArrayOnDevice) {
900+
if constexpr (!IsUsingArray) {
916901
m_Data = openclVector;
917902
} else {
918903
m_Data = bit_cast<DataType>(openclVector);
919904
}
920905
}
921906

922907
operator vector_t() const {
923-
if constexpr (!IsUsingArrayOnDevice) {
908+
if constexpr (!IsUsingArray) {
924909
return m_Data;
925910
} else {
926911
auto ptr = bit_cast<const VectorDataType *>((&m_Data)->data());
@@ -1081,7 +1066,7 @@ template <typename Type, int NumElements> class vec {
10811066
#ifdef __SYCL_DEVICE_ONLY__
10821067
#define __SYCL_BINOP(BINOP, OPASSIGN, CONVERT) \
10831068
template <typename Ty = vec> \
1084-
vec operator BINOP(const EnableIfNotUsingArrayOnDevice<Ty> &Rhs) const { \
1069+
vec operator BINOP(const EnableIfNotUsingArray<Ty> &Rhs) const { \
10851070
vec Ret; \
10861071
Ret.m_Data = m_Data BINOP Rhs.m_Data; \
10871072
if constexpr (std::is_same<Type, bool>::value && CONVERT) { \
@@ -1090,7 +1075,7 @@ template <typename Type, int NumElements> class vec {
10901075
return Ret; \
10911076
} \
10921077
template <typename Ty = vec> \
1093-
vec operator BINOP(const EnableIfUsingArrayOnDevice<Ty> &Rhs) const { \
1078+
vec operator BINOP(const EnableIfUsingArray<Ty> &Rhs) const { \
10941079
vec Ret; \
10951080
for (size_t I = 0; I < NumElements; ++I) { \
10961081
Ret.setValue(I, (getValue(I) BINOP Rhs.getValue(I))); \
@@ -1244,94 +1229,67 @@ template <typename Type, int NumElements> class vec {
12441229
__SYCL_UOP(--, -=)
12451230
#undef __SYCL_UOP
12461231

1247-
// operator~() available only when: dataT != float && dataT != double
1248-
// && dataT != half
1232+
// Available only when: dataT != cl_float && dataT != cl_double
1233+
// && dataT != cl_half
12491234
template <typename T = DataT>
1250-
typename std::enable_if_t<!std::is_floating_point_v<vec_data_t<T>> &&
1251-
(!IsUsingArrayOnDevice && !IsUsingArrayOnHost),
1252-
vec>
1235+
typename std::enable_if_t<std::is_integral_v<vec_data_t<T>>, vec>
12531236
operator~() const {
1237+
// Use __SYCL_DEVICE_ONLY__ macro because cast to OpenCL vector type is defined
1238+
// by SYCL device compiler only.
1239+
#ifdef __SYCL_DEVICE_ONLY__
12541240
vec Ret{(typename vec::DataType) ~m_Data};
12551241
if constexpr (std::is_same<Type, bool>::value) {
12561242
Ret.ConvertToDataT();
12571243
}
12581244
return Ret;
1259-
}
1260-
template <typename T = DataT>
1261-
typename std::enable_if_t<!std::is_floating_point_v<vec_data_t<T>> &&
1262-
(IsUsingArrayOnDevice || IsUsingArrayOnHost),
1263-
vec>
1264-
operator~() const {
1245+
#else
12651246
vec Ret{};
12661247
for (size_t I = 0; I < NumElements; ++I) {
12671248
Ret.setValue(I, ~getValue(I));
12681249
}
12691250
return Ret;
1251+
#endif
12701252
}
12711253

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

1318-
template <typename T = vec> EnableIfUsingArray<T> operator+() const {
1269+
vec operator+() const {
1270+
// Use __SYCL_DEVICE_ONLY__ macro because cast to OpenCL vector type is defined
1271+
// by SYCL device compiler only.
1272+
#ifdef __SYCL_DEVICE_ONLY__
1273+
return vec{+m_Data};
1274+
#else
13191275
vec Ret{};
13201276
for (size_t I = 0; I < NumElements; ++I)
13211277
Ret.setValue(I, vec_data<DataT>::get(+vec_data<DataT>::get(getValue(I))));
13221278
return Ret;
1279+
#endif
13231280
}
13241281

1325-
// operator -
1326-
template <typename T = vec> EnableIfNotUsingArray<T> operator-() const {
1282+
vec operator-() const {
1283+
// Use __SYCL_DEVICE_ONLY__ macro because cast to OpenCL vector type is defined
1284+
// by SYCL device compiler only.
1285+
#ifdef __SYCL_DEVICE_ONLY__
13271286
return vec{-m_Data};
1328-
}
1329-
1330-
template <typename T = vec> EnableIfUsingArray<T> operator-() const {
1287+
#else
13311288
vec Ret{};
13321289
for (size_t I = 0; I < NumElements; ++I)
13331290
Ret.setValue(I, vec_data<DataT>::get(-vec_data<DataT>::get(getValue(I))));
13341291
return Ret;
1292+
#endif
13351293
}
13361294

13371295
// OP is: &&, ||
@@ -1347,7 +1305,7 @@ template <typename Type, int NumElements> class vec {
13471305
template <template <typename> class Operation,
13481306
typename Ty = vec<DataT, NumElements>>
13491307
vec<DataT, NumElements>
1350-
operatorHelper(const EnableIfNotUsingArrayOnDevice<Ty> &Rhs) const {
1308+
operatorHelper(const EnableIfNotUsingArray<Ty> &Rhs) const {
13511309
vec<DataT, NumElements> Result;
13521310
Operation<DataType> Op;
13531311
Result.m_Data = Op(m_Data, Rhs.m_Data);
@@ -1357,7 +1315,7 @@ template <typename Type, int NumElements> class vec {
13571315
template <template <typename> class Operation,
13581316
typename Ty = vec<DataT, NumElements>>
13591317
vec<DataT, NumElements>
1360-
operatorHelper(const EnableIfUsingArrayOnDevice<Ty> &Rhs) const {
1318+
operatorHelper(const EnableIfUsingArray<Ty> &Rhs) const {
13611319
vec<DataT, NumElements> Result;
13621320
Operation<DataT> Op;
13631321
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)