Skip to content

Commit

Permalink
Fix clang-format issues.
Browse files Browse the repository at this point in the history
  • Loading branch information
@maarquitos14
maarquitos14 committed Nov 7, 2024
1 parent 320d1a8 commit 47ea34c
Showing 1 changed file with 51 additions and 50 deletions.
101 changes: 51 additions & 50 deletions sycl/test-e2e/LLVMIntrinsicLowering/intrinsic_cmp.cpp
View file
Original file line number Diff line number Diff line change
Expand Up @@ -18,26 +18,24 @@ typedef uint64_t v4u64_t __attribute__((ext_vector_type(4)));
// Check if a given type is a vector type or not. Used in submitAndCheck to
// branch the check: we need element-wise comparison for vector types. Default
// case: T is not a vector type.
template <typename T>
struct is_vector : std::false_type {};
template <typename T> struct is_vector : std::false_type {};
// Specialization for vector types. If T has
// __attribute__((ext_vector_type(N))), then it's a vector type.
template <typename T, std::size_t N>
struct is_vector<T __attribute__((ext_vector_type(N)))> : std::true_type {};
template <typename T>
inline constexpr bool is_vector_v = is_vector<T>::value;
template <typename T> inline constexpr bool is_vector_v = is_vector<T>::value;

// Get the length of a vector type. Used in submitAndCheck to iterate over the
// elements of the vector type. Default case: length is 1.
template <typename T>
struct vector_length {
static constexpr std::size_t value = 1;
static constexpr std::size_t value = 1;
};
// Specialization for vector types. If T has
// __attribute__((ext_vector_type(N))), then the length is N.
template <typename T, std::size_t N>
struct vector_length<T __attribute__((ext_vector_type(N)))> {
static constexpr std::size_t value = N;
static constexpr std::size_t value = N;
};
template <typename T>
inline constexpr std::size_t vector_length_v = vector_length<T>::value;
Expand All @@ -50,7 +48,7 @@ struct element_type;
// Specialization for vector types. If T has __attribute__((ext_vector_type(N))), return T.
template <typename T, int N>
struct element_type<T __attribute__((ext_vector_type(N)))> {
using type = T;
using type = T;
};
// Helper alias template.
template <typename T>
Expand All @@ -64,41 +62,41 @@ struct TypeList {};
// Function to trigger llvm.scmp/ucmp.
template <typename RetTy, typename ArgTy>
void compare(RetTy &res, ArgTy x, ArgTy y) {
auto lessOrEq = (x <= y);
auto lessThan = (x < y);
res = lessOrEq ? (lessThan ? RetTy(-1) : RetTy(0)) : RetTy(1);
auto lessOrEq = (x <= y);
auto lessThan = (x < y);
res = lessOrEq ? (lessThan ? RetTy(-1) : RetTy(0)) : RetTy(1);
}

// Function to submit kernel and check device result with host result.
template <typename RetTy, typename ArgTy>
void submitAndCheck(sycl::queue &q, ArgTy x, ArgTy y) {
RetTy res;
{
sycl::buffer<RetTy, 1> res_b{&res, 1};
q.submit([&](sycl::handler &cgh) {
sycl::accessor acc{res_b, cgh, sycl::write_only};
cgh.single_task<>([=] {
RetTy tmp;
compare<RetTy, ArgTy>(tmp, x, y);
acc[0] = tmp;
});
});
}
RetTy expectedRes;
compare<RetTy, ArgTy>(expectedRes, x, y);
if constexpr (is_vector_v<RetTy>) {
for (int i = 0; i < vector_length_v<RetTy>; ++i) {
assert(res[i] == expectedRes[i]);
}
} else {
assert(res == expectedRes);
RetTy res;
{
sycl::buffer<RetTy, 1> res_b{&res, 1};
q.submit([&](sycl::handler &cgh) {
sycl::accessor acc{res_b, cgh, sycl::write_only};
cgh.single_task<>([=] {
RetTy tmp;
compare<RetTy, ArgTy>(tmp, x, y);
acc[0] = tmp;
});
});
}
RetTy expectedRes;
compare<RetTy, ArgTy>(expectedRes, x, y);
if constexpr (is_vector_v<RetTy>) {
for (int i = 0; i < vector_length_v<RetTy>; ++i) {
assert(res[i] == expectedRes[i]);
}
} else {
assert(res == expectedRes);
}
}

// Helper to call submitAndCheck for each combination.
template <typename RetTypes, typename ArgTypes>
void submitAndCheckCombination(sycl::queue &q, int x, int y) {
submitAndCheck<RetTypes, ArgTypes>(q, x, y);
submitAndCheck<RetTypes, ArgTypes>(q, x, y);
}

// Function to generate all the combinations possible with the two type lists.
Expand All @@ -110,8 +108,8 @@ void submitAndCheckCombination(sycl::queue &q, int x, int y) {
// Recursive case to generate combinations.
template <typename RetType, typename... RetTypes, typename... ArgTypes>
void submitCombinations(sycl::queue &q, int x, int y, TypeList<RetType, RetTypes...>, TypeList<ArgTypes...>) {
(submitAndCheckCombination<RetType, ArgTypes>(q, x, y), ...);
submitCombinations(q, x, y, TypeList<RetTypes...>{}, TypeList<ArgTypes...>{});
(submitAndCheckCombination<RetType, ArgTypes>(q, x, y), ...);
submitCombinations(q, x, y, TypeList<RetTypes...>{}, TypeList<ArgTypes...>{});
}
// Base case to stop recursion.
template <typename... ArgTypes>
Expand All @@ -125,26 +123,29 @@ void submitCombinations(sycl::queue &, int, int, TypeList<>, TypeList<ArgTypes..
// Recursive case to generate combinations.
template <typename ArgType, typename... ArgTypes>
void submitVecCombinations(sycl::queue &q, int x, int y, TypeList<ArgType, ArgTypes...>) {
// Use signed types for return type, as it may return -1.
using ElemType = std::make_signed_t<element_type_t<ArgType>>;
using RetType = ElemType __attribute__((ext_vector_type(vector_length_v<ArgType>)));
submitAndCheckCombination<RetType, ArgType>(q, x, y);
submitVecCombinations(q, x, y, TypeList<ArgTypes...>{});
// Use signed types for return type, as it may return -1.
using ElemType = std::make_signed_t<element_type_t<ArgType>>;
using RetType =
ElemType __attribute__((ext_vector_type(vector_length_v<ArgType>)));
submitAndCheckCombination<RetType, ArgType>(q, x, y);
submitVecCombinations(q, x, y, TypeList<ArgTypes...>{});
}
// Base case to stop recursion.
void submitVecCombinations(sycl::queue &, int, int, TypeList<>) {}

int main(int argc, char **argv) {
sycl::queue q;
// RetTypes includes only signed types because it may return -1.
using RetTypes = TypeList<int8_t, int16_t, int32_t, int64_t>;
using ArgTypes = TypeList<int8_t, int16_t, int32_t, int64_t, uint8_t, uint16_t, uint32_t, uint64_t>;
submitCombinations(q, 50, 49, RetTypes{}, ArgTypes{});
submitCombinations(q, 50, 50, RetTypes{}, ArgTypes{});
submitCombinations(q, 50, 51, RetTypes{}, ArgTypes{});
using VecTypes = TypeList<v4i8_t, v4i16_t, v4i32_t, v4i64_t, v4u8_t, v4u16_t, v4u32_t, v4u64_t>;
submitVecCombinations(q, 50, 49, VecTypes{});
submitVecCombinations(q, 50, 50, VecTypes{});
submitVecCombinations(q, 50, 51, VecTypes{});
return 0;
sycl::queue q;
// RetTypes includes only signed types because it may return -1.
using RetTypes = TypeList<int8_t, int16_t, int32_t, int64_t>;
using ArgTypes = TypeList<int8_t, int16_t, int32_t, int64_t, uint8_t,
uint16_t, uint32_t, uint64_t>;
submitCombinations(q, 50, 49, RetTypes{}, ArgTypes{});
submitCombinations(q, 50, 50, RetTypes{}, ArgTypes{});
submitCombinations(q, 50, 51, RetTypes{}, ArgTypes{});
using VecTypes = TypeList<v4i8_t, v4i16_t, v4i32_t, v4i64_t, v4u8_t, v4u16_t,
v4u32_t, v4u64_t>;
submitVecCombinations(q, 50, 49, VecTypes{});
submitVecCombinations(q, 50, 50, VecTypes{});
submitVecCombinations(q, 50, 51, VecTypes{});
return 0;
}

0 comments on commit 47ea34c

Please sign in to comment.