Skip to content

Commit

Permalink
apply new clang format
Browse files Browse the repository at this point in the history
  • Loading branch information
AmirIpma committed Jun 28, 2023
1 parent 1ab967f commit ad634b6
Showing 1 changed file with 29 additions and 41 deletions.
70 changes: 29 additions & 41 deletions tests/common/invoke.h
Original file line number Diff line number Diff line change
Expand Up @@ -18,7 +18,7 @@ namespace {
* @brief Functor to invoke kernels with nd_item in use
* @tparam dim Dimension to use
* @tparam kernelT Type to use as the kernel name
*/
*/
template <int dim, typename kernelT>
struct invoke_nd_item {
static constexpr int dimensions = dim;
Expand All @@ -33,26 +33,23 @@ struct invoke_nd_item {
* @param kernelBody Kernel body to call
*/
template <typename kernelBodyT>
void operator()(sycl::handler &cgh,
sycl::range<dim> numWorkItems,
sycl::range<dim> workGroupSize,
kernelBodyT kernelBody) {

cgh.parallel_for<kernelT>(
sycl::nd_range<dim>(numWorkItems, workGroupSize),
[=](sycl::nd_item<dim> ndItem) {
const size_t index = ndItem.get_global_linear_id();

kernelBody(ndItem, index);
});
void operator()(sycl::handler& cgh, sycl::range<dim> numWorkItems,
sycl::range<dim> workGroupSize, kernelBodyT kernelBody) {
cgh.parallel_for<kernelT>(sycl::nd_range<dim>(numWorkItems, workGroupSize),
[=](sycl::nd_item<dim> ndItem) {
const size_t index =
ndItem.get_global_linear_id();

kernelBody(ndItem, index);
});
}
};

/**
* @brief Functor to invoke kernels with group in use
* @tparam dim Dimension to use
* @tparam kernelT Type to use as the kernel name
*/
*/
template <int dim, typename kernelT>
struct invoke_group {
static constexpr int dimensions = dim;
Expand All @@ -67,19 +64,16 @@ struct invoke_group {
* @param kernelBody Kernel body to call
*/
template <typename kernelBodyT>
void operator()(sycl::handler &cgh,
sycl::range<dim> numWorkItems,
sycl::range<dim> workGroupSize,
kernelBodyT kernelBody) {
void operator()(sycl::handler& cgh, sycl::range<dim> numWorkItems,
sycl::range<dim> workGroupSize, kernelBodyT kernelBody) {
sycl::range<dim> numWorkGroups = numWorkItems / workGroupSize;

cgh.parallel_for_work_group<kernelT>(
numWorkGroups, workGroupSize,
[=](sycl::group<dim> group) {
const size_t index = group.get_linear_id();
numWorkGroups, workGroupSize, [=](sycl::group<dim> group) {
const size_t index = group.get_linear_id();

kernelBody(group, index);
});
kernelBody(group, index);
});
}
};

Expand All @@ -88,11 +82,9 @@ struct invoke_sub_group {
static constexpr int dimensions = dim;
using parameterT = sycl::sub_group;

template <typename kernelBodyT>
void operator()(sycl::handler &cgh,
sycl::range<dim> numWorkItems,
sycl::range<dim> workGroupSize,
kernelBodyT kernelBody) {
template <typename kernelBodyT>
void operator()(sycl::handler& cgh, sycl::range<dim> numWorkItems,
sycl::range<dim> workGroupSize, kernelBodyT kernelBody) {
sycl::range<dim> numWorkGroups = numWorkItems / workGroupSize;

cgh.parallel_for<kernelT>(
Expand All @@ -102,19 +94,19 @@ template <typename kernelBodyT>
sycl::sub_group sub_group = item.get_sub_group();

kernelBody(sub_group, index);
});
});
}
};

/**
* @brief Generate and store the given number of nd_item/group/h_item/sub_group instances
* @brief Generate and store the given number of nd_item/group/h_item/sub_group
* instances
* @retval Array of instances
* @tparam numItems Number of instances to store
* @tparam kernelInvokeT Invocation functor to use
*/
template <size_t numItems, class kernelInvokeT>
std::array<typename kernelInvokeT::parameterT, numItems> store_instances()
{
std::array<typename kernelInvokeT::parameterT, numItems> store_instances() {
constexpr auto numDims = kernelInvokeT::dimensions;
using item_t = typename kernelInvokeT::parameterT;
using item_array_t = std::array<item_t, numItems>;
Expand All @@ -127,25 +119,21 @@ std::array<typename kernelInvokeT::parameterT, numItems> store_instances()
sycl_cts::util::get_cts_object::range<numDims>::get(numItems, 1, 1);

{
sycl::buffer<item_t> itemBuf(items.data(),
sycl::range<1>(items.size()));
sycl::buffer<item_t> itemBuf(items.data(), sycl::range<1>(items.size()));

auto queue = sycl_cts::util::get_cts_object::queue();
queue.submit([&](sycl::handler& cgh) {
auto itemAcc =
itemBuf.template get_access<sycl::access_mode::write>(cgh);
auto itemAcc = itemBuf.template get_access<sycl::access_mode::write>(cgh);

kernelInvokeT{}(
cgh, itemRange, oneElemRange,
[=](item_t& item, const size_t index) {
itemAcc[index] = item;
});
[=](item_t& item, const size_t index) { itemAcc[index] = item; });
});
queue.wait_and_throw();
}
return items;
}

} // namespace
} // namespace

#endif // __SYCLCTS_TESTS_COMMON_INVOKE_H
#endif // __SYCLCTS_TESTS_COMMON_INVOKE_H

0 comments on commit ad634b6

Please sign in to comment.