Skip to content

Uninitialized refactors #2549

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

Merged
danhoeflinger merged 19 commits into from
Mar 11, 2026
Merged

Uninitialized refactors #2549

Show file tree
Hide file tree
Changes from all commits
Commits
Show all changes
19 commits
Select commit Hold shift + click to select a range
67cd47b
refactor away redundant fill functor;
danhoeflinger Dec 19, 2025
7f178b0
fixing typo uninitialized_value_construct
danhoeflinger Dec 31, 2025
724fb32
revert formatting only changes
danhoeflinger Dec 31, 2025
62db354
fixing rebase
danhoeflinger Feb 20, 2026
516b995
formatting
danhoeflinger Feb 20, 2026
c315942
formatting
danhoeflinger Feb 20, 2026
fc3af7f
formatting
danhoeflinger Feb 25, 2026
a9f4a9a
move from algorithm to memory
danhoeflinger Feb 25, 2026
39972a9
reverting formatting only change
danhoeflinger Feb 25, 2026
76cdfd3
formatting fixes
danhoeflinger Feb 25, 2026
0f229d5
formatting
danhoeflinger Feb 25, 2026
ed00490
adding memory_impl_hetero.h
danhoeflinger Feb 25, 2026
36f57e0
formatting, naming
danhoeflinger Feb 25, 2026
5364964
fix header
danhoeflinger Feb 25, 2026
47cffba
remove stale sycl_traits
danhoeflinger Feb 25, 2026
89aec8e
using fully qualified name
danhoeflinger Feb 25, 2026
0a56ec7
fix header copyright
danhoeflinger Mar 2, 2026
2a2da89
fix no_init property
danhoeflinger Mar 6, 2026
3fcbd84
fully qualified name
danhoeflinger Mar 11, 2026
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
2 changes: 1 addition & 1 deletion include/oneapi/dpl/internal/async_impl/async_impl_hetero.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -172,7 +172,7 @@ __pattern_fill_async(__hetero_tag<_BackendTag> __tag, _ExecutionPolicy&& __exec,
__tag, ::std::forward<_ExecutionPolicy>(__exec),
__par_backend_hetero::make_iter_mode<__par_backend_hetero::access_mode::write>(__first),
__par_backend_hetero::make_iter_mode<__par_backend_hetero::access_mode::write>(__last),
fill_functor<_T>{__value});
oneapi::dpl::__internal::__brick_fill<__hetero_tag<_BackendTag>, _T>{__value});
}

//------------------------------------------------------------------------
Expand Down
17 changes: 0 additions & 17 deletions include/oneapi/dpl/pstl/algorithm_fwd.h
View file
Open in desktop
Comment thread
akukanov marked this conversation as resolved.
Original file line number Diff line number Diff line change
Expand Up @@ -74,15 +74,6 @@ template <class _IsVector, class _ExecutionPolicy, class _RandomAccessIterator,
void
__pattern_walk1(__parallel_tag<_IsVector>, _ExecutionPolicy&&, _RandomAccessIterator, _RandomAccessIterator, _Function);

template <class _Tag, class _ExecutionPolicy, class _ForwardIterator, class _Brick>
void
__pattern_walk_brick(_Tag, _ExecutionPolicy&&, _ForwardIterator, _ForwardIterator, _Brick) noexcept;

template <class _IsVector, class _ExecutionPolicy, class _RandomAccessIterator, class _Brick>
void
__pattern_walk_brick(__parallel_tag<_IsVector>, _ExecutionPolicy&&, _RandomAccessIterator, _RandomAccessIterator,
_Brick);

//------------------------------------------------------------------------
// walk1_n
//------------------------------------------------------------------------
Expand All @@ -103,14 +94,6 @@ template <class _IsVector, class _ExecutionPolicy, class _RandomAccessIterator,
_RandomAccessIterator
__pattern_walk1_n(__parallel_tag<_IsVector>, _ExecutionPolicy&&, _RandomAccessIterator, _Size, _Function);

template <class _Tag, class _ExecutionPolicy, class _ForwardIterator, class _Size, class _Brick>
_ForwardIterator
__pattern_walk_brick_n(_Tag, _ExecutionPolicy&&, _ForwardIterator, _Size, _Brick) noexcept;

template <class _IsVector, class _ExecutionPolicy, class _RandomAccessIterator, class _Size, class _Brick>
_RandomAccessIterator
__pattern_walk_brick_n(__parallel_tag<_IsVector>, _ExecutionPolicy&&, _RandomAccessIterator, _Size, _Brick);

//------------------------------------------------------------------------
// walk2 (pseudo)
//
Expand Down
48 changes: 0 additions & 48 deletions include/oneapi/dpl/pstl/algorithm_impl.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -177,30 +177,6 @@ __pattern_walk1(__parallel_tag<_IsVector>, _ExecutionPolicy&& __exec, _RandomAcc
});
}

template <class _Tag, class _ExecutionPolicy, class _ForwardIterator, class _Brick>
void
__pattern_walk_brick(_Tag, _ExecutionPolicy&&, _ForwardIterator __first, _ForwardIterator __last,
_Brick __brick) noexcept
{
static_assert(__is_serial_tag_v<_Tag> || __is_parallel_forward_tag_v<_Tag>);

__brick(__first, __last, typename _Tag::__is_vector{});
}

template <class _IsVector, class _ExecutionPolicy, class _RandomAccessIterator, class _Brick>
void
__pattern_walk_brick(__parallel_tag<_IsVector>, _ExecutionPolicy&& __exec, _RandomAccessIterator __first,
_RandomAccessIterator __last, _Brick __brick)
{
using __backend_tag = typename __parallel_tag<_IsVector>::__backend_tag;

__internal::__except_handler([&]() {
__par_backend::__parallel_for(
__backend_tag{}, ::std::forward<_ExecutionPolicy>(__exec), __first, __last,
[__brick](_RandomAccessIterator __i, _RandomAccessIterator __j) { __brick(__i, __j, _IsVector{}); });
});
}

//------------------------------------------------------------------------
// walk1_n
//------------------------------------------------------------------------
Expand Down Expand Up @@ -239,30 +215,6 @@ __pattern_walk1_n(__parallel_tag<_IsVector> __tag, _ExecutionPolicy&& __exec, _R
return __first + __n;
}

template <class _Tag, class _ExecutionPolicy, class _ForwardIterator, class _Size, class _Brick>
_ForwardIterator
__pattern_walk_brick_n(_Tag, _ExecutionPolicy&&, _ForwardIterator __first, _Size __n, _Brick __brick) noexcept
{
static_assert(__is_serial_tag_v<_Tag> || __is_parallel_forward_tag_v<_Tag>);

return __brick(__first, __n, typename _Tag::__is_vector{});
}

template <class _IsVector, class _ExecutionPolicy, class _RandomAccessIterator, class _Size, class _Brick>
_RandomAccessIterator
__pattern_walk_brick_n(__parallel_tag<_IsVector>, _ExecutionPolicy&& __exec, _RandomAccessIterator __first, _Size __n,
_Brick __brick)
{
using __backend_tag = typename __parallel_tag<_IsVector>::__backend_tag;

return __internal::__except_handler([&]() {
__par_backend::__parallel_for(
__backend_tag{}, ::std::forward<_ExecutionPolicy>(__exec), __first, __first + __n,
[__brick](_RandomAccessIterator __i, _RandomAccessIterator __j) { __brick(__i, __j - __i, _IsVector{}); });
return __first + __n;
});
}

//------------------------------------------------------------------------
// walk2 (pseudo)
//
Expand Down
33 changes: 15 additions & 18 deletions include/oneapi/dpl/pstl/glue_memory_impl.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -21,6 +21,7 @@

#if _ONEDPL_HETERO_BACKEND
# include "hetero/algorithm_impl_hetero.h"
# include "hetero/memory_impl_hetero.h"
#endif

#include "memory_fwd.h"
Expand Down Expand Up @@ -147,13 +148,12 @@ uninitialized_fill(_ExecutionPolicy&& __exec, _ForwardIterator __first, _Forward

if constexpr (oneapi::dpl::__internal::__trivial_uninitialized_fill<_ValueType, _Tp>)
{
oneapi::dpl::__internal::__pattern_walk_brick(
__dispatch_tag, std::forward<_ExecutionPolicy>(__exec), __first, __last,
oneapi::dpl::__internal::__brick_fill<decltype(__dispatch_tag), _ValueType>{_ValueType(__value)});
oneapi::dpl::__internal::__pattern_fill(__dispatch_tag, std::forward<_ExecutionPolicy>(__exec), __first, __last,
_ValueType(__value));
}
else
{
oneapi::dpl::__internal::__pattern_walk1(
oneapi::dpl::__internal::__pattern_uninitialized_walk1(
__dispatch_tag, std::forward<_ExecutionPolicy>(__exec), __first, __last,
oneapi::dpl::__internal::__op_uninitialized_fill<_Tp, std::decay_t<_ExecutionPolicy>>{__value});
}
Expand All @@ -169,13 +169,12 @@ uninitialized_fill_n(_ExecutionPolicy&& __exec, _ForwardIterator __first, _Size

if constexpr (oneapi::dpl::__internal::__trivial_uninitialized_fill<_ValueType, _Tp>)
{
return oneapi::dpl::__internal::__pattern_walk_brick_n(
__dispatch_tag, std::forward<_ExecutionPolicy>(__exec), __first, __n,
oneapi::dpl::__internal::__brick_fill_n<decltype(__dispatch_tag), _ValueType>{_ValueType(__value)});
return oneapi::dpl::__internal::__pattern_fill_n(__dispatch_tag, std::forward<_ExecutionPolicy>(__exec),
__first, __n, _ValueType(__value));
}
else
{
return oneapi::dpl::__internal::__pattern_walk1_n(
return oneapi::dpl::__internal::__pattern_uninitialized_walk1_n(
__dispatch_tag, std::forward<_ExecutionPolicy>(__exec), __first, __n,
oneapi::dpl::__internal::__op_uninitialized_fill<_Tp, std::decay_t<_ExecutionPolicy>>{__value});
}
Expand Down Expand Up @@ -243,7 +242,7 @@ uninitialized_default_construct(_ExecutionPolicy&& __exec, _ForwardIterator __fi
{
const auto __dispatch_tag = oneapi::dpl::__internal::__select_backend(__exec, __first);

oneapi::dpl::__internal::__pattern_walk1(
oneapi::dpl::__internal::__pattern_uninitialized_walk1(
__dispatch_tag, std::forward<_ExecutionPolicy>(__exec), __first, __last,
oneapi::dpl::__internal::__op_uninitialized_default_construct<std::decay_t<_ExecutionPolicy>>{});
}
Expand All @@ -263,7 +262,7 @@ uninitialized_default_construct_n(_ExecutionPolicy&& __exec, _ForwardIterator __
{
const auto __dispatch_tag = oneapi::dpl::__internal::__select_backend(__exec, __first);

return oneapi::dpl::__internal::__pattern_walk1_n(
return oneapi::dpl::__internal::__pattern_uninitialized_walk1_n(
__dispatch_tag, std::forward<_ExecutionPolicy>(__exec), __first, __n,
oneapi::dpl::__internal::__op_uninitialized_default_construct<std::decay_t<_ExecutionPolicy>>{});
}
Expand All @@ -281,13 +280,12 @@ uninitialized_value_construct(_ExecutionPolicy&& __exec, _ForwardIterator __firs

if constexpr (oneapi::dpl::__internal::__trivial_uninitialized_value_construct<_ValueType>)
{
oneapi::dpl::__internal::__pattern_walk_brick(
__dispatch_tag, std::forward<_ExecutionPolicy>(__exec), __first, __last,
oneapi::dpl::__internal::__brick_fill<decltype(__dispatch_tag), _ValueType>{_ValueType()});
oneapi::dpl::__internal::__pattern_fill(__dispatch_tag, std::forward<_ExecutionPolicy>(__exec), __first, __last,
_ValueType());
}
else
{
oneapi::dpl::__internal::__pattern_walk1(
oneapi::dpl::__internal::__pattern_uninitialized_walk1(
__dispatch_tag, std::forward<_ExecutionPolicy>(__exec), __first, __last,
oneapi::dpl::__internal::__op_uninitialized_value_construct<std::decay_t<_ExecutionPolicy>>{});
}
Expand All @@ -303,13 +301,12 @@ uninitialized_value_construct_n(_ExecutionPolicy&& __exec, _ForwardIterator __fi

if constexpr (oneapi::dpl::__internal::__trivial_uninitialized_value_construct<_ValueType>)
{
return oneapi::dpl::__internal::__pattern_walk_brick_n(
__dispatch_tag, std::forward<_ExecutionPolicy>(__exec), __first, __n,
oneapi::dpl::__internal::__brick_fill_n<decltype(__dispatch_tag), _ValueType>{_ValueType()});
return oneapi::dpl::__internal::__pattern_fill_n(__dispatch_tag, std::forward<_ExecutionPolicy>(__exec),
__first, __n, _ValueType());
}
else
{
return oneapi::dpl::__internal::__pattern_walk1_n(
return oneapi::dpl::__internal::__pattern_uninitialized_walk1_n(
__dispatch_tag, std::forward<_ExecutionPolicy>(__exec), __first, __n,
oneapi::dpl::__internal::__op_uninitialized_value_construct<std::decay_t<_ExecutionPolicy>>{});
}
Expand Down
51 changes: 1 addition & 50 deletions include/oneapi/dpl/pstl/hetero/algorithm_impl_hetero.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -214,43 +214,6 @@ __pattern_walk3(__hetero_tag<_BackendTag>, _ExecutionPolicy&& __exec, _ForwardIt
__f);
}

//------------------------------------------------------------------------
// walk_brick, walk_brick_n
//------------------------------------------------------------------------

template <typename _Name>
struct __walk_brick_wrapper;

template <typename _BackendTag, typename _ExecutionPolicy, typename _ForwardIterator, typename _Function>
void
__pattern_walk_brick(__hetero_tag<_BackendTag> __tag, _ExecutionPolicy&& __exec, _ForwardIterator __first,
_ForwardIterator __last, _Function __f)
{
if (__last - __first <= 0)
return;

__pattern_hetero_walk1<__par_backend_hetero::access_mode::read_write, false>(
__tag,
__par_backend_hetero::make_wrapped_policy<__walk_brick_wrapper>(::std::forward<_ExecutionPolicy>(__exec)),
Comment thread
SergeyKopienko marked this conversation as resolved.
__first, __last, __f);
}

template <typename _Name>
struct __walk_brick_n_wrapper;

template <typename _BackendTag, typename _ExecutionPolicy, typename _ForwardIterator, typename _Size,
typename _Function>
_ForwardIterator
__pattern_walk_brick_n(__hetero_tag<_BackendTag> __tag, _ExecutionPolicy&& __exec, _ForwardIterator __first, _Size __n,
_Function __f)
{
__pattern_hetero_walk1<__par_backend_hetero::access_mode::read_write, false>(
__tag,
__par_backend_hetero::make_wrapped_policy<__walk_brick_n_wrapper>(::std::forward<_ExecutionPolicy>(__exec)),
__first, __first + __n, __f);
return __first + __n;
}

//------------------------------------------------------------------------
// walk2_brick, walk2_brick_n
//------------------------------------------------------------------------
Expand Down Expand Up @@ -333,18 +296,6 @@ __pattern_walk3_transform_if(__hetero_tag<_BackendTag> __tag, _ExecutionPolicy&&
// fill
//------------------------------------------------------------------------

template <typename _SourceT>
struct fill_functor
{
_SourceT __value;
template <typename _TargetT>
void
operator()(_TargetT& __target) const
{
__target = __value;
}
};

template <typename _BackendTag, typename _ExecutionPolicy, typename _ForwardIterator, typename _T>
_ForwardIterator
__pattern_fill(__hetero_tag<_BackendTag> __tag, _ExecutionPolicy&& __exec, _ForwardIterator __first,
Expand All @@ -354,7 +305,7 @@ __pattern_fill(__hetero_tag<_BackendTag> __tag, _ExecutionPolicy&& __exec, _Forw
__tag, std::forward<_ExecutionPolicy>(__exec),
__par_backend_hetero::make_iter_mode<__par_backend_hetero::access_mode::write>(__first),
__par_backend_hetero::make_iter_mode<__par_backend_hetero::access_mode::write>(__last),
fill_functor<_T>{__value});
__brick_fill<__hetero_tag<_BackendTag>, _T>{__value});
return __last;
}

Expand Down
9 changes: 0 additions & 9 deletions include/oneapi/dpl/pstl/hetero/dpcpp/sycl_traits.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -95,9 +95,6 @@ class __replace_functor;
template <typename _Tp, typename _Pred>
class __replace_copy_functor;

template <typename _SourceT>
struct fill_functor;

template <typename _Generator>
struct generate_functor;

Expand Down Expand Up @@ -248,12 +245,6 @@ struct sycl::is_device_copyable<_ONEDPL_SPECIALIZE_FOR(oneapi::dpl::__internal::
{
};

template <typename _SourceT>
struct sycl::is_device_copyable<_ONEDPL_SPECIALIZE_FOR(oneapi::dpl::__internal::fill_functor, _SourceT)>
: oneapi::dpl::__internal::__are_all_device_copyable<_SourceT>
{
};

template <typename _Generator>
struct sycl::is_device_copyable<_ONEDPL_SPECIALIZE_FOR(oneapi::dpl::__internal::generate_functor, _Generator)>
: oneapi::dpl::__internal::__are_all_device_copyable<_Generator>
Expand Down
3 changes: 2 additions & 1 deletion include/oneapi/dpl/pstl/hetero/histogram_impl_hetero.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -134,7 +134,8 @@ __pattern_histogram(__hetero_tag<_BackendTag>, _ExecutionPolicy&& __exec, _Rando
auto __bins_buf = __keep_bins(__histogram_first, __histogram_first + __num_bins);
auto __bins = __bins_buf.all_view();

auto __fill_func = oneapi::dpl::__internal::fill_functor<_global_histogram_type>{_global_histogram_type{0}};
auto __fill_func = oneapi::dpl::__internal::__brick_fill<__hetero_tag<_BackendTag>, _global_histogram_type>{
_global_histogram_type{0}};
//fill histogram bins with zeros

auto __init_event = oneapi::dpl::__par_backend_hetero::__parallel_for(
Expand Down
54 changes: 54 additions & 0 deletions include/oneapi/dpl/pstl/hetero/memory_impl_hetero.h
View file
Open in desktop
Original file line number Diff line number Diff line change
@@ -0,0 +1,54 @@
// -*- C++ -*-
//===----------------------------------------------------------------------===//
//
// Copyright (C) UXL Foundation Contributors
//
// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
//
//===----------------------------------------------------------------------===//

#ifndef _ONEDPL_MEMORY_IMPL_HETERO_H
#define _ONEDPL_MEMORY_IMPL_HETERO_H

#include "algorithm_impl_hetero.h"

namespace oneapi
{
namespace dpl
{
namespace __internal
{

//------------------------------------------------------------------------
// uninitialized_walk1
//------------------------------------------------------------------------

template <class _BackendTag, class _ExecutionPolicy, class _ForwardIterator, class _Function>
void
__pattern_uninitialized_walk1(__hetero_tag<_BackendTag> tag, _ExecutionPolicy&& __exec, _ForwardIterator __first,
_ForwardIterator __last, _Function __f)
{
oneapi::dpl::__internal::__pattern_hetero_walk1<sycl::access_mode::write, /*_IsNoInitRequested=*/true>(
tag, std::forward<_ExecutionPolicy>(__exec), __first, __last, __f);
}

//------------------------------------------------------------------------
// uninitialized_walk1_n
//------------------------------------------------------------------------

template <typename _BackendTag, typename _ExecutionPolicy, typename _ForwardIterator, typename _Size,
typename _Function>
_ForwardIterator
__pattern_uninitialized_walk1_n(__hetero_tag<_BackendTag> __tag, _ExecutionPolicy&& __exec, _ForwardIterator __first,
_Size __n, _Function __f)
{
oneapi::dpl::__internal::__pattern_hetero_walk1<sycl::access_mode::write, /*_IsNoInitRequested=*/true>(
__tag, std::forward<_ExecutionPolicy>(__exec), __first, __first + __n, __f);
Comment thread
danhoeflinger marked this conversation as resolved.
return __first + __n;
}

} // namespace __internal
} // namespace dpl
} // namespace oneapi

#endif // _ONEDPL_MEMORY_IMPL_HETERO_H
8 changes: 8 additions & 0 deletions include/oneapi/dpl/pstl/memory_fwd.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -41,6 +41,14 @@ struct __op_uninitialized_default_construct;
template <typename... _ExecutionPolicy>
struct __op_uninitialized_value_construct;

template <class _Tag, class _ExecutionPolicy, class _ForwardIterator, class _Function>
void
__pattern_uninitialized_walk1(_Tag, _ExecutionPolicy&&, _ForwardIterator, _ForwardIterator, _Function);

template <class _Tag, class _ExecutionPolicy, class _ForwardIterator, class _Size, class _Function>
_ForwardIterator
__pattern_uninitialized_walk1_n(_Tag, _ExecutionPolicy&&, _ForwardIterator, _Size, _Function);

} // namespace __internal
} // namespace dpl
} // namespace oneapi
Expand Down
Loading
Loading