Skip to content

Commit

Permalink
Add constructors for for-based bricks
Browse files Browse the repository at this point in the history 
Signed-off-by: Matthew Michel <[email protected]>
  • Loading branch information
@mmichel11
mmichel11 committed Jan 4, 2025
1 parent fedd5de commit 08aa260
Show file tree
Hide file tree
Showing 5 changed files with 57 additions and 23 deletions.
6 changes: 3 additions & 3 deletions include/oneapi/dpl/internal/async_impl/async_impl_hetero.h
View file
Original file line number Diff line number Diff line change
Expand Up @@ -47,7 +47,7 @@ __pattern_walk1_async(__hetero_tag<_BackendTag>, _ExecutionPolicy&& __exec, _For
auto __future_obj = oneapi::dpl::__par_backend_hetero::__parallel_for(
_BackendTag{}, ::std::forward<_ExecutionPolicy>(__exec),
unseq_backend::walk1_vector_or_scalar<_ExecutionPolicy, _Function, decltype(__view)>{
{}, __f, static_cast<std::size_t>(__n)},
__f, static_cast<std::size_t>(__n)},
__n, __view);
return __future_obj;
}
Expand Down Expand Up @@ -75,7 +75,7 @@ __pattern_walk2_async(__hetero_tag<_BackendTag>, _ExecutionPolicy&& __exec, _For
auto __future = oneapi::dpl::__par_backend_hetero::__parallel_for(
_BackendTag{}, ::std::forward<_ExecutionPolicy>(__exec),
unseq_backend::walk2_vectors_or_scalars<_ExecutionPolicy, _Function, decltype(__view1), decltype(__view2)>{
{}, __f, static_cast<std::size_t>(__n)},
__f, static_cast<std::size_t>(__n)},
__n, __view1, __view2);

return __future.__make_future(__first2 + __n);
Expand Down Expand Up @@ -107,7 +107,7 @@ __pattern_walk3_async(__hetero_tag<_BackendTag>, _ExecutionPolicy&& __exec, _For
auto __future = oneapi::dpl::__par_backend_hetero::__parallel_for(
_BackendTag{}, ::std::forward<_ExecutionPolicy>(__exec),
unseq_backend::walk3_vectors_or_scalars<_ExecutionPolicy, _Function, decltype(__view1), decltype(__view2),
decltype(__view3)>{{}, __f, static_cast<size_t>(__n)},
decltype(__view3)>{__f, static_cast<size_t>(__n)},
__n, __view1, __view2, __view3);

return __future.__make_future(__first3 + __n);
Expand Down
22 changes: 11 additions & 11 deletions include/oneapi/dpl/pstl/hetero/algorithm_impl_hetero.h
View file
Original file line number Diff line number Diff line change
Expand Up @@ -60,7 +60,7 @@ __pattern_walk1(__hetero_tag<_BackendTag>, _ExecutionPolicy&& __exec, _ForwardIt
oneapi::dpl::__par_backend_hetero::__parallel_for(
_BackendTag{}, __exec,
unseq_backend::walk1_vector_or_scalar<_ExecutionPolicy, _Function, decltype(__view)>{
{}, __f, static_cast<std::size_t>(__n)},
__f, static_cast<std::size_t>(__n)},
__n, __view)
.__deferrable_wait();
}
Expand Down Expand Up @@ -111,7 +111,7 @@ __pattern_walk2(__hetero_tag<_BackendTag>, _ExecutionPolicy&& __exec, _ForwardIt
auto __future = oneapi::dpl::__par_backend_hetero::__parallel_for(
_BackendTag{}, ::std::forward<_ExecutionPolicy>(__exec),
unseq_backend::walk2_vectors_or_scalars<_ExecutionPolicy, _Function, decltype(__view1), decltype(__view2)>{
{}, __f, static_cast<std::size_t>(__n)},
__f, static_cast<std::size_t>(__n)},
__n, __view1, __view2);

// Call no wait, wait or deferrable wait depending on _WaitMode
Expand Down Expand Up @@ -157,7 +157,7 @@ __pattern_swap(__hetero_tag<_BackendTag> __tag, _ExecutionPolicy&& __exec, _Forw
auto __future = oneapi::dpl::__par_backend_hetero::__parallel_for(
_BackendTag{}, std::forward<_ExecutionPolicy>(__exec),
unseq_backend::__brick_swap<_ExecutionPolicy, _Function, decltype(__view1), decltype(__view2)>{
{}, __f, static_cast<std::size_t>(__n)},
__f, static_cast<std::size_t>(__n)},
__n, __view1, __view2);
__future.wait(__par_backend_hetero::__deferrable_mode{});
return __first2 + __n;
Expand Down Expand Up @@ -194,7 +194,7 @@ __pattern_walk3(__hetero_tag<_BackendTag>, _ExecutionPolicy&& __exec, _ForwardIt
oneapi::dpl::__par_backend_hetero::__parallel_for(
_BackendTag{}, std::forward<_ExecutionPolicy>(__exec),
unseq_backend::walk3_vectors_or_scalars<_ExecutionPolicy, _Function, decltype(__view1), decltype(__view2),
decltype(__view3)>{{}, __f, static_cast<std::size_t>(__n)},
decltype(__view3)>{__f, static_cast<std::size_t>(__n)},
__n, __view1, __view2, __view3)
.__deferrable_wait();

Expand Down Expand Up @@ -1599,7 +1599,7 @@ __pattern_reverse(__hetero_tag<_BackendTag>, _ExecutionPolicy&& __exec, _Iterato
oneapi::dpl::__par_backend_hetero::__parallel_for(
_BackendTag{}, std::forward<_ExecutionPolicy>(__exec),
unseq_backend::__reverse_functor<typename std::iterator_traits<_Iterator>::difference_type,
decltype(__buf.all_view())>{{}, __n},
decltype(__buf.all_view())>{__n},
__n / 2, __buf.all_view())
.__deferrable_wait();
}
Expand Down Expand Up @@ -1628,7 +1628,7 @@ __pattern_reverse_copy(__hetero_tag<_BackendTag>, _ExecutionPolicy&& __exec, _Bi
oneapi::dpl::__par_backend_hetero::__parallel_for(
_BackendTag{}, std::forward<_ExecutionPolicy>(__exec),
unseq_backend::__reverse_copy<typename std::iterator_traits<_BidirectionalIterator>::difference_type,
decltype(__view1), decltype(__view2)>{{}, __n},
decltype(__view1), decltype(__view2)>{__n},
__n, __view1, __view2)
.__deferrable_wait();

Expand Down Expand Up @@ -1671,7 +1671,7 @@ __pattern_rotate(__hetero_tag<_BackendTag>, _ExecutionPolicy&& __exec, _Iterator
oneapi::dpl::__par_backend_hetero::__parallel_for(
_BackendTag{}, oneapi::dpl::__par_backend_hetero::make_wrapped_policy<__rotate_wrapper>(__exec),
unseq_backend::__rotate_copy<typename std::iterator_traits<_Iterator>::difference_type, decltype(__view),
decltype(__temp_rng_w)>{{}, __n, __shift},
decltype(__temp_rng_w)>{__n, __shift},
__n, __view, __temp_rng_w);

//An explicit wait isn't required here because we are working with a temporary sycl::buffer and sycl accessors and
Expand All @@ -1680,9 +1680,9 @@ __pattern_rotate(__hetero_tag<_BackendTag>, _ExecutionPolicy&& __exec, _Iterator
using _Function = __brick_move<__hetero_tag<_BackendTag>, _ExecutionPolicy>;
auto __temp_rng_rw =
oneapi::dpl::__ranges::all_view<_Tp, __par_backend_hetero::access_mode::read_write>(__temp_buf.get_buffer());
auto __brick = unseq_backend::walk2_vectors_or_scalars<_ExecutionPolicy, _Function, decltype(__temp_rng_rw),
decltype(__buf.all_view())>{
{}, _Function{}, static_cast<std::size_t>(__n)};
auto __brick =
unseq_backend::walk2_vectors_or_scalars<_ExecutionPolicy, _Function, decltype(__temp_rng_rw),
decltype(__buf.all_view())>{_Function{}, static_cast<std::size_t>(__n)};
oneapi::dpl::__par_backend_hetero::__parallel_for(_BackendTag{}, std::forward<_ExecutionPolicy>(__exec), __brick,
__n, __temp_rng_rw, __buf.all_view())
.__deferrable_wait();
Expand Down Expand Up @@ -1721,7 +1721,7 @@ __pattern_rotate_copy(__hetero_tag<_BackendTag>, _ExecutionPolicy&& __exec, _Bid
oneapi::dpl::__par_backend_hetero::__parallel_for(
_BackendTag{}, ::std::forward<_ExecutionPolicy>(__exec),
unseq_backend::__rotate_copy<typename ::std::iterator_traits<_BidirectionalIterator>::difference_type,
decltype(__view1), decltype(__view2)>{{}, __n, __shift},
decltype(__view1), decltype(__view2)>{__n, __shift},
__n, __view1, __view2)
.__deferrable_wait();

Expand Down
15 changes: 7 additions & 8 deletions include/oneapi/dpl/pstl/hetero/algorithm_ranges_impl_hetero.h
View file
Original file line number Diff line number Diff line change
Expand Up @@ -64,7 +64,7 @@ __pattern_walk_n(__hetero_tag<_BackendTag>, _ExecutionPolicy&& __exec, _Function
oneapi::dpl::__par_backend_hetero::__parallel_for(
_BackendTag{}, ::std::forward<_ExecutionPolicy>(__exec),
unseq_backend::walk1_vector_or_scalar<_ExecutionPolicy, _Function, std::decay_t<_Ranges>...>{
{}, __f, static_cast<std::size_t>(__n)},
__f, static_cast<std::size_t>(__n)},
__n, ::std::forward<_Ranges>(__rngs)...)
.__deferrable_wait();
}
Expand All @@ -73,7 +73,7 @@ __pattern_walk_n(__hetero_tag<_BackendTag>, _ExecutionPolicy&& __exec, _Function
oneapi::dpl::__par_backend_hetero::__parallel_for(
_BackendTag{}, ::std::forward<_ExecutionPolicy>(__exec),
unseq_backend::walk2_vectors_or_scalars<_ExecutionPolicy, _Function, std::decay_t<_Ranges>...>{
{}, __f, static_cast<std::size_t>(__n)},
__f, static_cast<std::size_t>(__n)},
__n, ::std::forward<_Ranges>(__rngs)...)
.__deferrable_wait();
}
Expand All @@ -82,7 +82,7 @@ __pattern_walk_n(__hetero_tag<_BackendTag>, _ExecutionPolicy&& __exec, _Function
oneapi::dpl::__par_backend_hetero::__parallel_for(
_BackendTag{}, ::std::forward<_ExecutionPolicy>(__exec),
unseq_backend::walk3_vectors_or_scalars<_ExecutionPolicy, _Function, std::decay_t<_Ranges>...>{
{}, __f, static_cast<std::size_t>(__n)},
__f, static_cast<std::size_t>(__n)},
__n, ::std::forward<_Ranges>(__rngs)...)
.__deferrable_wait();
}
Expand Down Expand Up @@ -178,7 +178,7 @@ __pattern_swap(__hetero_tag<_BackendTag> __tag, _ExecutionPolicy&& __exec, _Rang
auto __future = oneapi::dpl::__par_backend_hetero::__parallel_for(
_BackendTag{}, std::move(__exec1),
unseq_backend::__brick_swap<decltype(__exec1), _Function, std::decay_t<_Range1>, std::decay_t<_Range2>>{
{}, __f, __n},
__f, __n},
__n, __rng1, __rng2);
__future.wait(__par_backend_hetero::__deferrable_mode{});
return __n;
Expand All @@ -188,8 +188,8 @@ __pattern_swap(__hetero_tag<_BackendTag> __tag, _ExecutionPolicy&& __exec, _Rang
oneapi::dpl::__par_backend_hetero::make_wrapped_policy<__swap2_wrapper>(std::forward<_ExecutionPolicy>(__exec));
auto __future = oneapi::dpl::__par_backend_hetero::__parallel_for(
_BackendTag{}, std::move(__exec2),
unseq_backend::__brick_swap<decltype(__exec2), _Function, std::decay_t<_Range2>, std::decay_t<_Range1>>{
{}, __f, __n},
unseq_backend::__brick_swap<decltype(__exec2), _Function, std::decay_t<_Range2>, std::decay_t<_Range1>>{__f,
__n},
__n, __rng2, __rng1);
__future.wait(__par_backend_hetero::__deferrable_mode{});
return __n;
Expand Down Expand Up @@ -658,8 +658,7 @@ __pattern_unique_copy(__hetero_tag<_BackendTag> __tag, _ExecutionPolicy&& __exec
oneapi::dpl::__par_backend_hetero::make_wrapped_policy<__copy_wrapper>(
std::forward<_ExecutionPolicy>(__exec)),
unseq_backend::walk2_vectors_or_scalars<_ExecutionPolicy, _CopyBrick, std::decay_t<_Range1>,
std::decay_t<_Range2>>{
{}, _CopyBrick{}, static_cast<std::size_t>(__n)},
std::decay_t<_Range2>>{_CopyBrick{}, static_cast<std::size_t>(__n)},
__n, std::forward<_Range1>(__rng), std::forward<_Range2>(__result))
.get();

Expand Down
35 changes: 35 additions & 0 deletions include/oneapi/dpl/pstl/hetero/dpcpp/unseq_backend_sycl.h
View file
Original file line number Diff line number Diff line change
Expand Up @@ -155,10 +155,14 @@ struct walk_scalar_base
template <typename _ExecutionPolicy, typename _F, typename _Range>
struct walk1_vector_or_scalar : public walk_vector_or_scalar_base<_Range>
{
private:
using __base_t = walk_vector_or_scalar_base<_Range>;
_F __f;
std::size_t __n;

public:
walk1_vector_or_scalar(_F __f, std::size_t __n) : __f(__f), __n(__n) {}

template <typename _IsFull, typename _ItemId>
void
__vector_path(_IsFull __is_full, const _ItemId __idx, _Range __rng) const
Expand Down Expand Up @@ -191,10 +195,14 @@ struct walk1_vector_or_scalar : public walk_vector_or_scalar_base<_Range>
template <typename _ExecutionPolicy, typename _F, typename _Range1, typename _Range2>
struct walk2_vectors_or_scalars : public walk_vector_or_scalar_base<_Range1, _Range2>
{
private:
using __base_t = walk_vector_or_scalar_base<_Range1, _Range2>;
_F __f;
std::size_t __n;

public:
walk2_vectors_or_scalars(_F __f, std::size_t __n) : __f(__f), __n(__n) {}

template <typename _IsFull, typename _ItemId>
void
__vector_path(_IsFull __is_full, const _ItemId __idx, _Range1 __rng1, _Range2 __rng2) const
Expand Down Expand Up @@ -242,10 +250,14 @@ struct walk2_vectors_or_scalars : public walk_vector_or_scalar_base<_Range1, _Ra
template <typename _ExecutionPolicy, typename _F, typename _Range1, typename _Range2, typename _Range3>
struct walk3_vectors_or_scalars : public walk_vector_or_scalar_base<_Range1, _Range2, _Range3>
{
private:
using __base_t = walk_vector_or_scalar_base<_Range1, _Range2, _Range3>;
_F __f;
std::size_t __n;

public:
walk3_vectors_or_scalars(_F __f, std::size_t __n) : __f(__f), __n(__n) {}

template <typename _IsFull, typename _ItemId>
void
__vector_path(_IsFull __is_full, const _ItemId __idx, _Range1 __rng1, _Range2 __rng2, _Range3 __rng3) const
Expand Down Expand Up @@ -322,10 +334,14 @@ struct walk_n<_ExecutionPolicy, oneapi::dpl::__internal::__no_op>
template <typename _ExecutionPolicy, typename _F, typename _Range1, typename _Range2>
struct walk_adjacent_difference : public walk_vector_or_scalar_base<_Range1, _Range2>
{
private:
using __base_t = walk_vector_or_scalar_base<_Range1, _Range2>;
_F __f;
std::size_t __n;

public:
walk_adjacent_difference(_F __f, std::size_t __n) : __f(__f), __n(__n) {}

template <typename _IsFull, typename _ItemId>
void
__scalar_path(_IsFull, const _ItemId __idx, const _Range1 __rng1, _Range2 __rng2) const
Expand Down Expand Up @@ -1156,9 +1172,14 @@ struct __brick_includes
template <typename _Size, typename _Range>
struct __reverse_functor : public walk_vector_or_scalar_base<_Range>
{
private:
using __base_t = walk_vector_or_scalar_base<_Range>;
using _ValueType = oneapi::dpl::__internal::__value_t<_Range>;
_Size __size;

public:
__reverse_functor(_Size __size) : __size(__size) {}

template <typename _IsFull, typename _Idx>
void
__vector_path(_IsFull __is_full, const _Idx __left_start_idx, _Range __rng) const
Expand Down Expand Up @@ -1231,10 +1252,14 @@ struct __reverse_functor : public walk_vector_or_scalar_base<_Range>
template <typename _Size, typename _Range1, typename _Range2>
struct __reverse_copy : public walk_vector_or_scalar_base<_Range1, _Range2>
{
private:
using __base_t = walk_vector_or_scalar_base<_Range1, _Range2>;
using _ValueType = oneapi::dpl::__internal::__value_t<_Range1>;
_Size __size;

public:
__reverse_copy(_Size __size) : __size(__size) {}

template <typename _IsFull, typename _Idx>
void
__scalar_path(_IsFull, const _Idx __idx, const _Range1 __rng1, _Range2 __rng2) const
Expand Down Expand Up @@ -1296,10 +1321,15 @@ struct __reverse_copy : public walk_vector_or_scalar_base<_Range1, _Range2>
template <typename _Size, typename _Range1, typename _Range2>
struct __rotate_copy : public walk_vector_or_scalar_base<_Range1, _Range2>
{
private:
using __base_t = walk_vector_or_scalar_base<_Range1, _Range2>;
using _ValueType = oneapi::dpl::__internal::__value_t<_Range1>;
_Size __size;
_Size __shift;

public:
__rotate_copy(_Size __size, _Size __shift) : __size(__size), __shift(__shift) {}

template <typename _IsFull, typename _Idx>
void
__vector_path(_IsFull __is_full, const _Idx __idx, const _Range1 __rng1, _Range2 __rng2) const
Expand Down Expand Up @@ -1515,9 +1545,14 @@ struct __brick_reduce_idx : public walk_scalar_base<_Range>
template <typename _ExecutionPolicy, typename _F, typename _Range1, typename _Range2>
struct __brick_swap : public walk_vector_or_scalar_base<_Range1, _Range2>
{
private:
using __base_t = walk_vector_or_scalar_base<_Range1, _Range2>;
_F __f;
std::size_t __n;

public:
__brick_swap(_F __f, std::size_t __n) : __f(__f), __n(__n) {}

template <typename _IsFull, typename _ItemId>
void
__vector_path(_IsFull __is_full, const _ItemId __idx, _Range1 __rng1, _Range2 __rng2) const
Expand Down
2 changes: 1 addition & 1 deletion include/oneapi/dpl/pstl/hetero/numeric_impl_hetero.h
View file
Original file line number Diff line number Diff line change
Expand Up @@ -270,7 +270,7 @@ __pattern_adjacent_difference(__hetero_tag<_BackendTag> __tag, _ExecutionPolicy&
decltype(__view2)>;

oneapi::dpl::__par_backend_hetero::__parallel_for(
_BackendTag{}, __exec, _Function{{}, __fn, static_cast<std::size_t>(__n)}, __n, __view1, __view2)
_BackendTag{}, __exec, _Function{__fn, static_cast<std::size_t>(__n)}, __n, __view1, __view2)
.__deferrable_wait();
}

Expand Down

0 comments on commit 08aa260

Please sign in to comment.