Skip to content

Commit

Permalink
Merge branch 'sycl-upstream' into maxime/enable-imm-cmd-list-tests
Browse files Browse the repository at this point in the history
  • Loading branch information
mfrancepillois committed Feb 9, 2024
2 parents bf38afe + c200fe3 commit 6024463
Show file tree
Hide file tree
Showing 9 changed files with 181 additions and 10 deletions.
9 changes: 8 additions & 1 deletion sycl/include/sycl/builtins_preview.hpp
Original file line number Diff line number Diff line change
Expand Up @@ -137,7 +137,14 @@ auto builtin_marray_impl(FuncTy F, const Ts &...x) {
marray<ret_elem_type, T::size()> Res;
constexpr auto N = T::size();
for (size_t I = 0; I < N / 2; ++I) {
auto PartialRes = F(to_vec2(x, I * 2)...);
auto PartialRes = [&]() {
using elem_ty = get_elem_type_t<T>;
if constexpr (std::is_integral_v<elem_ty>)
return F(to_vec2(x, I * 2)
.template as<vec<get_fixed_sized_int_t<elem_ty>, 2>>()...);
else
return F(to_vec2(x, I * 2)...);
}();
std::memcpy(&Res[I * 2], &PartialRes, sizeof(decltype(PartialRes)));
}
if (N % 2)
Expand Down
11 changes: 11 additions & 0 deletions sycl/include/sycl/builtins_utils_scalar.hpp
Original file line number Diff line number Diff line change
Expand Up @@ -128,6 +128,17 @@ template <size_t Size> struct get_unsigned_int_by_size {
template <typename T> struct same_size_unsigned_int {
using type = typename get_unsigned_int_by_size<sizeof(T)>::type;
};
template <typename T>
using same_size_unsigned_int_t = typename same_size_unsigned_int<T>::type;

template <typename T> struct get_fixed_sized_int {
static_assert(std::is_integral_v<T>);
using type =
std::conditional_t<std::is_signed_v<T>, same_size_signed_int_t<T>,
same_size_unsigned_int_t<T>>;
};
template <typename T>
using get_fixed_sized_int_t = typename get_fixed_sized_int<T>::type;

// Utility trait for getting an upsampled integer type.
// NOTE: For upsampling we look for an integer of double the size of the
Expand Down
8 changes: 8 additions & 0 deletions sycl/include/sycl/detail/builtins/helper_macros.hpp
Original file line number Diff line number Diff line change
Expand Up @@ -48,6 +48,11 @@
FOR_EACH4_A6(BASE_CASE, FIXED1, FIXED2, FIXED3, FIXED4, ARG1, ARG2, ARG3, \
ARG4, ARG5, ARG6) \
BASE_CASE(FIXED1, FIXED2, FIXED3, FIXED4, ARG7)
#define FOR_EACH4_A8(BASE_CASE, FIXED1, FIXED2, FIXED3, FIXED4, ARG1, ARG2, \
ARG3, ARG4, ARG5, ARG6, ARG7, ARG8) \
FOR_EACH4_A7(BASE_CASE, FIXED1, FIXED2, FIXED3, FIXED4, ARG1, ARG2, ARG3, \
ARG4, ARG5, ARG6, ARG7) \
BASE_CASE(FIXED1, FIXED2, FIXED3, FIXED4, ARG8)
#define FOR_EACH4_A11(BASE_CASE, FIXED1, FIXED2, FIXED3, FIXED4, ARG1, ARG2, \
ARG3, ARG4, ARG5, ARG6, ARG7, ARG8, ARG9, ARG10, ARG11) \
FOR_EACH4_A7(BASE_CASE, FIXED1, FIXED2, FIXED3, FIXED4, ARG1, ARG2, ARG3, \
Expand Down Expand Up @@ -169,6 +174,9 @@
unsigned char, unsigned short, unsigned int, unsigned long, unsigned long long
// 11 types
#define INTEGER_TYPES SIGNED_TYPES, UNSIGNED_TYPES
// 8 types
#define FIXED_WIDTH_INTEGER_TYPES \
int8_t, int16_t, int32_t, int64_t, uint8_t, uint16_t, uint32_t, uint64_t

#define DEVICE_IMPL_TEMPLATE_CUSTOM_DELEGATE( \
NUM_ARGS, NAME, ENABLER, DELEGATOR, NS, /*SCALAR_VEC_IMPL*/...) \
Expand Down
7 changes: 4 additions & 3 deletions sycl/include/sycl/detail/builtins/integer_functions.inc
Original file line number Diff line number Diff line change
Expand Up @@ -16,9 +16,10 @@ namespace detail {
template <typename T>
struct integer_elem_type
: std::bool_constant<
check_type_in_v<get_elem_type_t<T>, char, signed char, short, int,
long, long long, unsigned char, unsigned short,
unsigned int, unsigned long, unsigned long long>> {};
(is_vec_or_swizzle_v<T> &&
check_type_in_v<get_elem_type_t<T>, FIXED_WIDTH_INTEGER_TYPES>) ||
(!is_vec_or_swizzle_v<T> &&
check_type_in_v<get_elem_type_t<T>, INTEGER_TYPES>)> {};
template <typename T>
struct suint32_elem_type
: std::bool_constant<
Expand Down
10 changes: 6 additions & 4 deletions sycl/include/sycl/detail/builtins/relational_functions.inc
Original file line number Diff line number Diff line change
Expand Up @@ -15,10 +15,12 @@ inline namespace _V1 {
namespace detail {
template <typename T>
struct bitselect_elem_type
: std::bool_constant<check_type_in_v<
get_elem_type_t<T>, float, double, half, char, signed char, short,
int, long, long long, unsigned char, unsigned short, unsigned int,
unsigned long, unsigned long long>> {};
: std::bool_constant<
check_type_in_v<get_elem_type_t<T>, FP_TYPES> ||
(is_vec_or_swizzle_v<T> &&
check_type_in_v<get_elem_type_t<T>, FIXED_WIDTH_INTEGER_TYPES>) ||
(!is_vec_or_swizzle_v<T> &&
check_type_in_v<get_elem_type_t<T>, INTEGER_TYPES>)> {};

template <typename T>
struct rel_ret_traits
Expand Down
7 changes: 7 additions & 0 deletions sycl/source/builtins/host_helper_macros.hpp
Original file line number Diff line number Diff line change
Expand Up @@ -56,6 +56,9 @@
#define EXPORT_VEC(NUM_ARGS, NAME, TYPE, VL) \
EXPORT_VEC_NS(NUM_ARGS, NAME, sycl, TYPE, VL)

#define EXPORT_VEC_1_16_IMPL(NUM_ARGS, NAME, NS, TYPE) \
FOR_VEC_1_16(EXPORT_VEC_NS, NUM_ARGS, NAME, NS, TYPE)

#define EXPORT_SCALAR_AND_VEC_1_16_IMPL(NUM_ARGS, NAME, NS, TYPE) \
EXPORT_SCALAR_NS(NUM_ARGS, NAME, NS, TYPE) \
FOR_VEC_1_16(EXPORT_VEC_NS, NUM_ARGS, NAME, NS, TYPE)
Expand All @@ -69,8 +72,12 @@

#define EXPORT_SCALAR_AND_VEC_1_16_NS(NUM_ARGS, NAME, NS, ...) \
FOR_EACH3(EXPORT_SCALAR_AND_VEC_1_16_IMPL, NUM_ARGS, NAME, NS, __VA_ARGS__)
#define EXPORT_VEC_1_16_NS(NUM_ARGS, NAME, NS, ...) \
FOR_EACH3(EXPORT_VEC_1_16_IMPL, NUM_ARGS, NAME, NS, __VA_ARGS__)
#define EXPORT_SCALAR_AND_VEC_1_16(NUM_ARGS, NAME, ...) \
EXPORT_SCALAR_AND_VEC_1_16_NS(NUM_ARGS, NAME, sycl, __VA_ARGS__)
#define EXPORT_VEC_1_16(NUM_ARGS, NAME, ...) \
EXPORT_VEC_1_16_NS(NUM_ARGS, NAME, sycl, __VA_ARGS__)

#define EXPORT_SCALAR_AND_VEC_2_4(NUM_ARGS, NAME, ...) \
FOR_EACH2(EXPORT_SCALAR_AND_VEC_2_4_IMPL, NUM_ARGS, NAME, __VA_ARGS__)
Expand Down
3 changes: 2 additions & 1 deletion sycl/source/builtins/integer_functions.cpp
Original file line number Diff line number Diff line change
Expand Up @@ -76,7 +76,8 @@ namespace sycl {
inline namespace _V1 {
#define BUILTIN_GENINT(NUM_ARGS, NAME, IMPL) \
HOST_IMPL(NAME, IMPL) \
EXPORT_SCALAR_AND_VEC_1_16(NUM_ARGS, NAME, INTEGER_TYPES)
FOR_EACH2(EXPORT_SCALAR, NUM_ARGS, NAME, INTEGER_TYPES) \
EXPORT_VEC_1_16(NUM_ARGS, NAME, FIXED_WIDTH_INTEGER_TYPES)
#define BUILTIN_GENINT_SU(NUM_ARGS, NAME, IMPL) \
BUILTIN_GENINT(NUM_ARGS, NAME, IMPL)

Expand Down
3 changes: 2 additions & 1 deletion sycl/source/builtins/relational_functions.cpp
Original file line number Diff line number Diff line change
Expand Up @@ -103,6 +103,7 @@ HOST_IMPL(bitselect, [](auto x, auto y, auto z) {
assert((ures & std::numeric_limits<utype>::max()) == ures);
return bit_cast<T0>(static_cast<utype>(ures));
})
EXPORT_SCALAR_AND_VEC_1_16(THREE_ARGS, bitselect, INTEGER_TYPES, FP_TYPES)
FOR_EACH2(EXPORT_SCALAR, THREE_ARGS, bitselect, INTEGER_TYPES, FP_TYPES)
EXPORT_VEC_1_16(THREE_ARGS, bitselect, FIXED_WIDTH_INTEGER_TYPES, FP_TYPES)
} // namespace _V1
} // namespace sycl
133 changes: 133 additions & 0 deletions sycl/test/basic_tests/builtins/builtin_unit_tests.cpp
Original file line number Diff line number Diff line change
@@ -0,0 +1,133 @@
// RUN: %clangxx -fsycl -fpreview-breaking-changes -fsyntax-only %s -Xclang -verify
// REQUIRES: preview-breaking-changes-supported

#include <sycl/sycl.hpp>

using namespace sycl;
using namespace sycl::detail;

namespace builtin_same_shape_v_tests {
using swizzle1 = decltype(std::declval<vec<float, 2>>().swizzle<0>());
using swizzle2 = decltype(std::declval<vec<float, 2>>().swizzle<0, 0>());
using swizzle3 = decltype(std::declval<vec<float, 2>>().swizzle<0, 0, 1>());

static_assert(builtin_same_shape_v<float>);
static_assert(builtin_same_shape_v<int, float>);
static_assert(builtin_same_shape_v<marray<int, 2>>);
static_assert(builtin_same_shape_v<marray<int, 2>, marray<float, 2>>);
static_assert(builtin_same_shape_v<vec<int, 2>>);
static_assert(builtin_same_shape_v<vec<int, 2>, vec<float, 2>>);
static_assert(builtin_same_shape_v<vec<int, 2>, swizzle2>);

static_assert(!builtin_same_shape_v<float, marray<float, 1>>);
static_assert(!builtin_same_shape_v<float, vec<float, 1>>);
static_assert(!builtin_same_shape_v<marray<float, 1>, vec<float, 1>>);
static_assert(!builtin_same_shape_v<float, swizzle1>);
static_assert(!builtin_same_shape_v<marray<float, 1>, swizzle1>);
static_assert(!builtin_same_shape_v<swizzle2, swizzle1>);
} // namespace builtin_same_shape_v_tests

namespace builtin_marray_impl_tests {
// Integer functions/relational bitselect only accept fixed-width integer
// element types for vector/swizzle elements. Make sure that our marray->vec
// delegator can handle that.

auto foo(char x) { return x; }
auto foo(signed char x) { return x; }
auto foo(unsigned char x) { return x; }
auto foo(vec<int8_t, 2> x) { return x; }
auto foo(vec<uint8_t, 2> x) { return x; }

auto test() {
marray<char, 2> x;
marray<signed char, 2> y;
marray<unsigned char, 2> z;
auto TestOne = [](auto x) {
std::ignore = builtin_marray_impl([](auto x) { return foo(x); }, x);
};
TestOne(x);
TestOne(y);
TestOne(z);
}
} // namespace builtin_marray_impl_tests

namespace builtin_enable_integer_tests {
using swizzle1 = decltype(std::declval<vec<int8_t, 2>>().swizzle<0>());
using swizzle2 = decltype(std::declval<vec<int8_t, 2>>().swizzle<0, 0>());
template <typename... Ts> void ignore() {}

void test() {
// clang-format off
ignore<builtin_enable_integer_t<char>,
builtin_enable_integer_t<signed char>,
builtin_enable_integer_t<unsigned char>>();
// clang-format on

ignore<builtin_enable_integer_t<vec<int8_t, 2>>,
builtin_enable_integer_t<vec<uint8_t, 2>>>();

ignore<builtin_enable_integer_t<char, char>>();
ignore<builtin_enable_integer_t<vec<int8_t, 2>, vec<int8_t, 2>>>();
ignore<builtin_enable_integer_t<vec<int8_t, 2>, swizzle2>>();
ignore<builtin_enable_integer_t<swizzle2, swizzle2>>();

{
// Only one of char/signed char maps onto int8_t. The other type isn't a
// valid vector element type for integer builtins.

static_assert(std::is_signed_v<char>);

// clang-format off
// expected-error-re@*:* {{no type named 'type' in 'sycl::detail::builtin_enable<sycl::detail::default_ret_type, sycl::detail::integer_elem_type, sycl::detail::any_shape, sycl::detail::same_elem_type, sycl::vec<{{.*}}, 2>>'}}
// expected-note@+1 {{in instantiation of template type alias 'builtin_enable_integer_t' requested here}}
ignore<builtin_enable_integer_t<vec<signed char, 2>>, builtin_enable_integer_t<vec<char, 2>>>();
// clang-format on
}

// expected-error@*:* {{no type named 'type' in 'sycl::detail::builtin_enable<sycl::detail::default_ret_type, sycl::detail::integer_elem_type, sycl::detail::any_shape, sycl::detail::same_elem_type, char, signed char>'}}
// expected-note@+1 {{in instantiation of template type alias 'builtin_enable_integer_t' requested here}}
ignore<builtin_enable_integer_t<char, signed char>>();
}
} // namespace builtin_enable_integer_tests

namespace builtin_enable_bitselect_tests {
// Essentially the same as builtin_enable_integer_t + FP types support.
using swizzle1 = decltype(std::declval<vec<int8_t, 2>>().swizzle<0>());
using swizzle2 = decltype(std::declval<vec<int8_t, 2>>().swizzle<0, 0>());
template <typename... Ts> void ignore() {}

void test() {
// clang-format off
ignore<builtin_enable_bitselect_t<char>,
builtin_enable_bitselect_t<signed char>,
builtin_enable_bitselect_t<unsigned char>,
builtin_enable_bitselect_t<float>>();
// clang-format on

ignore<builtin_enable_bitselect_t<vec<int8_t, 2>>,
builtin_enable_bitselect_t<vec<uint8_t, 2>>,
builtin_enable_bitselect_t<vec<float, 2>>>();

ignore<builtin_enable_bitselect_t<char, char>>();
ignore<builtin_enable_bitselect_t<vec<int8_t, 2>, vec<int8_t, 2>>>();
ignore<builtin_enable_bitselect_t<vec<int8_t, 2>, swizzle2>>();
ignore<builtin_enable_bitselect_t<swizzle2, swizzle2>>();

{
// Only one of char/signed char maps onto int8_t. The other type isn't a
// valid vector element type for integer builtins.

static_assert(std::is_signed_v<char>);

// clang-format off
// expected-error-re@*:* {{no type named 'type' in 'sycl::detail::builtin_enable<sycl::detail::default_ret_type, sycl::detail::bitselect_elem_type, sycl::detail::any_shape, sycl::detail::same_elem_type, sycl::vec<{{.*}}, 2>>'}}
// expected-note@+1 {{in instantiation of template type alias 'builtin_enable_bitselect_t' requested here}}
ignore<builtin_enable_bitselect_t<vec<signed char, 2>>, builtin_enable_bitselect_t<vec<char, 2>>>();
// clang-format on
}

// expected-error@*:* {{no type named 'type' in 'sycl::detail::builtin_enable<sycl::detail::default_ret_type, sycl::detail::bitselect_elem_type, sycl::detail::any_shape, sycl::detail::same_elem_type, char, signed char>'}}
// expected-note@+1 {{in instantiation of template type alias 'builtin_enable_bitselect_t' requested here}}
ignore<builtin_enable_bitselect_t<char, signed char>>();
}
} // namespace builtin_enable_bitselect_tests

0 comments on commit 6024463

Please sign in to comment.