Skip to content

[SYCL] Make vec constexpr constructor work with half #8189

New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

Sign up for GitHub

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Closed
wants to merge 1 commit into from
Closed

[SYCL] Make vec constexpr constructor work with half #8189

Show file tree
Hide file tree
Changes from all commits
Commits
Show all changes
1 commit
Select commit
File filter

Filter by extension

Filter by extension
Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
3 changes: 3 additions & 0 deletions sycl/include/sycl/half_type.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -252,6 +252,9 @@ template <int NumElements> struct half_vec {
StorageT s[NumElements];

__SYCL_CONSTEXPR_HALF half_vec() : s{0.0f} { initialize_data(); }
// This constructor is needed by vec class, see description of
// MDataInitializerType there.
__SYCL_CONSTEXPR_HALF half_vec(float) : s{0.0f} { initialize_data(); }
constexpr void initialize_data() {
for (size_t i = 0; i < NumElements; ++i) {
s[i] = StorageT(0.0f);
Expand Down
37 changes: 33 additions & 4 deletions sycl/include/sycl/types.hpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -563,6 +563,31 @@ template <typename Type, int NumElements> class vec {
using DataType =
typename detail::BaseCLTypeConverter<DataT, NumElements>::DataType;

// Before C++20, it is necessary for constexpr constructor to have all
// non-static class members initializers, which means that 'DataT m_Data' will
// be initialized by a default constexpr constructor in thi case.
//
// Unfortunately, we don't have constexpr default constructor for sycl::half.
// It is not possible to have both trivial and constexpr default constructor
// before C++20 and we decided that it is better to have trivial default
// constructor for sycl::half. See intel/llvm#4631. SYCL spec is of no help
// here, because half is underspecified there, see KhronosGroup/SYCL-Docs#350.
//
// To not require default constexpr constructor from underlying DataType we
// instead do the following trick in "affected" constructors:
// vec(...) : m_Data{ static_cast<MDataInitializerType>(0) } { ... }
// i.e. we rely on non-default constexpr constructor of DataType.
//
// Since amount of types, which can be used as underlying type of vec is
// limited, we consider it safe to assume that they have the required
// non-default constexpr constructor.
//
// Strictly speaking, DataType is not the same as DataT and if we can't use
// ext_vector_type attribute, then it could be some SSE/AVX extension type.
// FIXME: Are we sure that they all can be constructed in a way we want?
using MDataInitializerType =
std::conditional_t<detail::is_floating_point<DataT>::value, float, int>;

static constexpr int getNumElements() { return NumElements; }

// SizeChecker is needed for vec(const argTN &... args) ctor to validate args.
Expand Down Expand Up @@ -712,7 +737,8 @@ template <typename Type, int NumElements> class vec {
T>;

template <typename Ty = DataT>
explicit constexpr vec(const EnableIfNotHostHalf<Ty> &arg) {
explicit constexpr vec(const EnableIfNotHostHalf<Ty> &arg)
: m_Data{static_cast<MDataInitializerType>(0)} {
m_Data = (DataType)vec_data<Ty>::get(arg);
}

Expand All @@ -727,7 +753,8 @@ template <typename Type, int NumElements> class vec {
}

template <typename Ty = DataT>
explicit constexpr vec(const EnableIfHostHalf<Ty> &arg) {
explicit constexpr vec(const EnableIfHostHalf<Ty> &arg)
: m_Data{static_cast<MDataInitializerType>(0)} {
for (int i = 0; i < NumElements; ++i) {
setValue(i, arg);
}
Expand All @@ -745,7 +772,8 @@ template <typename Type, int NumElements> class vec {
return *this;
}
#else
explicit constexpr vec(const DataT &arg) {
explicit constexpr vec(const DataT &arg)
: m_Data{static_cast<MDataInitializerType>(0)} {
for (int i = 0; i < NumElements; ++i) {
setValue(i, arg);
}
Expand Down Expand Up @@ -819,7 +847,8 @@ template <typename Type, int NumElements> class vec {
// base types are match and that the NumElements == sum of lengths of args.
template <typename... argTN, typename = EnableIfSuitableTypes<argTN...>,
typename = EnableIfSuitableNumElements<argTN...>>
constexpr vec(const argTN &...args) {
constexpr vec(const argTN &...args)
: m_Data{static_cast<MDataInitializerType>(0)} {
vaargCtorHelper(0, args...);
}

Expand Down
65 changes: 65 additions & 0 deletions sycl/test/basic_tests/vectors/constexpr-constructor.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
@@ -0,0 +1,65 @@
// RUN: %clangxx -fsycl -fsyntax-only -Wno-deprecated-declarations %s
// RUN: %clangxx -fsycl -D__NO_EXT_VECTOR_TYPE_ON_HOST__ -fsyntax-only -Wno-deprecated-declarations %s

#include <sycl/sycl.hpp>

#include <cstdint>

#define DEFINE_CONSTEXPR_VECTOR(name, type, size) \
constexpr sycl::vec<type, size> name##_##size{0};

#define DEFINE_CONSTEXPR_VECTOR_INIT_NON_ZERO(name, type, size, init) \
constexpr sycl::vec<type, size> name##_##size{init};

#define DEFINE_CONSTEXPR_VECTOR_FOR_TYPE(type) \
DEFINE_CONSTEXPR_VECTOR(type, type, 1)\
DEFINE_CONSTEXPR_VECTOR(type, type, 2)\
DEFINE_CONSTEXPR_VECTOR(type, type, 3)\
DEFINE_CONSTEXPR_VECTOR(type, type, 4)\
DEFINE_CONSTEXPR_VECTOR(type, type, 8)\
DEFINE_CONSTEXPR_VECTOR(type, type, 16)

#define DEFINE_CONSTEXPR_VECTOR_FOR_TYPE_NAMED(type, name) \
DEFINE_CONSTEXPR_VECTOR(name, type, 1)\
DEFINE_CONSTEXPR_VECTOR(name, type, 2)\
DEFINE_CONSTEXPR_VECTOR(name, type, 3)\
DEFINE_CONSTEXPR_VECTOR(name, type, 4)\
DEFINE_CONSTEXPR_VECTOR(name, type, 8)\
DEFINE_CONSTEXPR_VECTOR(name, type, 16)

int main() {

DEFINE_CONSTEXPR_VECTOR_FOR_TYPE_NAMED(sycl::byte, syclbyte)
DEFINE_CONSTEXPR_VECTOR_FOR_TYPE_NAMED(sycl::half, half)
DEFINE_CONSTEXPR_VECTOR_FOR_TYPE(bool)
DEFINE_CONSTEXPR_VECTOR_FOR_TYPE(char)
DEFINE_CONSTEXPR_VECTOR_FOR_TYPE_NAMED(signed char, schar)
DEFINE_CONSTEXPR_VECTOR_FOR_TYPE_NAMED(unsigned char, uchar)
DEFINE_CONSTEXPR_VECTOR_FOR_TYPE_NAMED(short int, short)
DEFINE_CONSTEXPR_VECTOR_FOR_TYPE_NAMED(unsigned short int, ushort)
DEFINE_CONSTEXPR_VECTOR_FOR_TYPE(int)
DEFINE_CONSTEXPR_VECTOR_FOR_TYPE_NAMED(unsigned int, uint)
DEFINE_CONSTEXPR_VECTOR_FOR_TYPE_NAMED(long int, long)
DEFINE_CONSTEXPR_VECTOR_FOR_TYPE_NAMED(unsigned long int, ulong)
DEFINE_CONSTEXPR_VECTOR_FOR_TYPE_NAMED(long long int, longlong)
DEFINE_CONSTEXPR_VECTOR_FOR_TYPE_NAMED(unsigned long long int, ulonglong)
DEFINE_CONSTEXPR_VECTOR_FOR_TYPE(float)
DEFINE_CONSTEXPR_VECTOR_FOR_TYPE(double)
DEFINE_CONSTEXPR_VECTOR_FOR_TYPE_NAMED(std::int8_t, int8)
DEFINE_CONSTEXPR_VECTOR_FOR_TYPE_NAMED(std::uint8_t, uint8)
DEFINE_CONSTEXPR_VECTOR_FOR_TYPE_NAMED(std::int16_t, int16)
DEFINE_CONSTEXPR_VECTOR_FOR_TYPE_NAMED(std::uint16_t, uint16)
DEFINE_CONSTEXPR_VECTOR_FOR_TYPE_NAMED(std::int32_t, int32)
DEFINE_CONSTEXPR_VECTOR_FOR_TYPE_NAMED(std::uint32_t, uint32)
DEFINE_CONSTEXPR_VECTOR_FOR_TYPE_NAMED(std::int64_t, int64)
DEFINE_CONSTEXPR_VECTOR_FOR_TYPE_NAMED(std::uint64_t, uint64)

DEFINE_CONSTEXPR_VECTOR_INIT_NON_ZERO(stdbyte, std::byte, 1, std::byte{1});
DEFINE_CONSTEXPR_VECTOR_INIT_NON_ZERO(stdbyte, std::byte, 2, std::byte{1});
DEFINE_CONSTEXPR_VECTOR_INIT_NON_ZERO(stdbyte, std::byte, 3, std::byte{1});
DEFINE_CONSTEXPR_VECTOR_INIT_NON_ZERO(stdbyte, std::byte, 4, std::byte{1});
DEFINE_CONSTEXPR_VECTOR_INIT_NON_ZERO(stdbyte, std::byte, 8, std::byte{1});
DEFINE_CONSTEXPR_VECTOR_INIT_NON_ZERO(stdbyte, std::byte, 16, std::byte{1});

return 0;
}