non-trivially-copyable comparators aren't supported for sort/sort_by_key

[masterleinad]

Describe the Bug: Migrated from https://github.com/kokkos/kokkos/issues/6939.

Compiling

#include <sycl/sycl.hpp>

#include <oneapi/dpl/algorithm>
#include <oneapi/dpl/execution>
#include <cstdio>
#include <iostream>

// When the special members are defined, we get the following error
//#define HIDE_SPECIAL_MEMBERS

// User-defined comparator we don't have control over, that does not specialize sycl::is_device_copyable
// Obviously, this functor is in fact device-copyable but that is not the case in general.
struct UserComparator{

  bool operator()(int i, int j)const { return keys[i] < keys[j];}

  int* keys;
};

// Our attempt to guarantee that functor argument passed to oneDPL sort is device-copyable
template <typename Functor, typename ValueType>
class CompareWrapper {
  Functor m_functor;

 public:
  bool operator()(const ValueType& lhs, const ValueType& rhs) const {
      return m_functor(lhs, rhs);
  }

    CompareWrapper(const Functor& f) : m_functor(f) {}

#ifndef HIDE_SPECIAL_MEMBERS
    CompareWrapper(const CompareWrapper& other) {
      std::memcpy(&m_functor, &other.m_functor, sizeof(m_functor));
    }
    CompareWrapper(CompareWrapper&& other) {
      std::memcpy(&m_functor, &other.m_functor, sizeof(m_functor));
    }
    CompareWrapper& operator=(const CompareWrapper& other) {
      std::memcpy(&m_functor, &other.m_functor, sizeof(m_functor));
      return *this;
    }
    CompareWrapper& operator=(CompareWrapper&& other) {
      std::memcpy(&m_functor, &other.m_functor, sizeof(m_functor));
      return *this;
    }
    ~CompareWrapper(){};
#endif
};

#ifndef HIDE_SPECIAL_MEMBERS
template<typename Functor, typename ValueType>
struct
sycl::is_device_copyable<CompareWrapper<Functor, ValueType>> : std::true_type {};
#endif

int main(int argc, char* argv[]) {
  sycl::queue queue;

  const int n = 10;

  int* values_ptr = sycl::malloc_device<int>(n, queue);
  int* keys_ptr = sycl::malloc_device<int>(n, queue);

  UserComparator user_comparator{keys_ptr};
  CompareWrapper<UserComparator, int> comparator{user_comparator};
  auto policy = oneapi::dpl::execution::make_device_policy(queue);
  static_assert(sycl::is_device_copyable_v<decltype(comparator)>);
  oneapi::dpl::sort(policy, values_ptr, values_ptr + n, comparator);
}

via

icpx -fsycl -I ~/onedpl-install/include/ test_Onedpl.cpp

results in

In file included from test_Onedpl.cpp:1:
In file included from /soft/preview/pe/24.180.0-RC5/CNDA/oneapi/compiler/eng-20240629/bin/compiler/../../include/sycl/sycl.hpp:25:
In file included from /soft/preview/pe/24.180.0-RC5/CNDA/oneapi/compiler/eng-20240629/bin/compiler/../../include/sycl/detail/core.hpp:21:
In file included from /soft/preview/pe/24.180.0-RC5/CNDA/oneapi/compiler/eng-20240629/bin/compiler/../../include/sycl/accessor.hpp:15:
In file included from /soft/preview/pe/24.180.0-RC5/CNDA/oneapi/compiler/eng-20240629/bin/compiler/../../include/sycl/buffer.hpp:19:
/soft/preview/pe/24.180.0-RC5/CNDA/oneapi/compiler/eng-20240629/bin/compiler/../../include/sycl/detail/is_device_copyable.hpp:109:17: error: static assertion failed due to requirement 'is_device_copyable_v<oneapi::dpl::__par_backend_hetero::__leaf_sorter<oneapi::dpl::__ranges::guard_view<int *>, oneapi::dpl::__internal::__compare<CompareWrapper<UserComparator, int>, oneapi::dpl::identity>>> || detail::IsDeprecatedDeviceCopyable<oneapi::dpl::__par_backend_hetero::__leaf_sorter<oneapi::dpl::__ranges::guard_view<int *>, oneapi::dpl::__internal::__compare<CompareWrapper<UserComparator, int>, oneapi::dpl::identity>>, void>::value': The specified type is not device copyable
  109 |   static_assert(is_device_copyable_v<FieldT> ||
      |                 ^~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~
  110 |                     detail::IsDeprecatedDeviceCopyable<FieldT>::value,
      |                     ~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~
/soft/preview/pe/24.180.0-RC5/CNDA/oneapi/compiler/eng-20240629/bin/compiler/../../include/sycl/detail/is_device_copyable.hpp:107:7: note: in instantiation of template class 'sycl::detail::CheckFieldsAreDeviceCopyable<(lambda at /home/darndt/onedpl-install/include/oneapi/dpl/pstl/hetero/../hetero/dpcpp/parallel_backend_sycl_merge_sort.h:234:29), 1>' requested here
  107 |     : CheckFieldsAreDeviceCopyable<T, NumFieldsToCheck - 1> {
      |       ^
/soft/preview/pe/24.180.0-RC5/CNDA/oneapi/compiler/eng-20240629/bin/compiler/../../include/sycl/detail/is_device_copyable.hpp:142:7: note: in instantiation of template class 'sycl::detail::CheckFieldsAreDeviceCopyable<(lambda at /home/darndt/onedpl-install/include/oneapi/dpl/pstl/hetero/../hetero/dpcpp/parallel_backend_sycl_merge_sort.h:234:29), 2>' requested here
  142 |     : CheckFieldsAreDeviceCopyable<FuncT, __builtin_num_fields(FuncT)>,
      |       ^
/soft/preview/pe/24.180.0-RC5/CNDA/oneapi/compiler/eng-20240629/bin/compiler/../../include/sycl/handler.hpp:1765:5: note: in instantiation of template class 'sycl::detail::CheckDeviceCopyable<(lambda at /home/darndt/onedpl-install/include/oneapi/dpl/pstl/hetero/../hetero/dpcpp/parallel_backend_sycl_merge_sort.h:234:29)>' requested here
 1765 |     detail::CheckDeviceCopyable<KernelType>();
      |     ^
/soft/preview/pe/24.180.0-RC5/CNDA/oneapi/compiler/eng-20240629/bin/compiler/../../include/sycl/handler.hpp:1807:5: note: in instantiation of function template specialization 'sycl::handler::unpack<(lambda at /home/darndt/onedpl-install/include/oneapi/dpl/pstl/hetero/../hetero/dpcpp/parallel_backend_sycl_merge_sort.h:234:29), (lambda at /home/darndt/onedpl-install/include/oneapi/dpl/pstl/hetero/../hetero/dpcpp/parallel_backend_sycl_merge_sort.h:234:29), sycl::ext::oneapi::experimental::properties<std::tuple<>>, false, (lambda at /soft/preview/pe/24.180.0-RC5/CNDA/oneapi/compiler/eng-20240629/bin/compiler/../../include/sycl/handler.hpp:1810:21)>' requested here
 1807 |     unpack<KernelName, KernelType, PropertiesT,
      |     ^
/soft/preview/pe/24.180.0-RC5/CNDA/oneapi/compiler/eng-20240629/bin/compiler/../../include/sycl/handler.hpp:1480:5: note: in instantiation of function template specialization 'sycl::handler::kernel_parallel_for_wrapper<(lambda at /home/darndt/onedpl-install/include/oneapi/dpl/pstl/hetero/../hetero/dpcpp/parallel_backend_sycl_merge_sort.h:234:29), sycl::nd_item<>, (lambda at /home/darndt/onedpl-install/include/oneapi/dpl/pstl/hetero/../hetero/dpcpp/parallel_backend_sycl_merge_sort.h:234:29), sycl::ext::oneapi::experimental::properties<std::tuple<>>>' requested here
 1480 |     kernel_parallel_for_wrapper<NameT, TransformedArgType, KernelType,
      |     ^
/soft/preview/pe/24.180.0-RC5/CNDA/oneapi/compiler/eng-20240629/bin/compiler/../../include/sycl/handler.hpp:2506:5: note: (skipping 1 context in backtrace; use -ftemplate-backtrace-limit=0 to see all)
 2506 |     parallel_for_impl<KernelName>(Range, Properties, std::move(KernelFunc));
      |     ^
/soft/preview/pe/24.180.0-RC5/CNDA/oneapi/compiler/eng-20240629/bin/compiler/../../include/sycl/handler.hpp:2591:5: note: in instantiation of function template specialization 'sycl::handler::parallel_for<sycl::detail::auto_name, (lambda at /home/darndt/onedpl-install/include/oneapi/dpl/pstl/hetero/../hetero/dpcpp/parallel_backend_sycl_merge_sort.h:234:29), sycl::ext::oneapi::experimental::properties<std::tuple<>>, 1>' requested here
 2591 |     parallel_for<KernelName>(Range,
      |     ^
/home/darndt/onedpl-install/include/oneapi/dpl/pstl/hetero/../hetero/dpcpp/parallel_backend_sycl_merge_sort.h:233:19: note: in instantiation of function template specialization 'sycl::handler::parallel_for<sycl::detail::auto_name, 1, (lambda at /home/darndt/onedpl-install/include/oneapi/dpl/pstl/hetero/../hetero/dpcpp/parallel_backend_sycl_merge_sort.h:234:29)>' requested here
  233 |             __cgh.parallel_for<_LeafSortName...>(
      |                   ^
/home/darndt/onedpl-install/include/oneapi/dpl/pstl/hetero/algorithm_impl_hetero.h:1249:5: note: in instantiation of function template specialization 'oneapi::dpl::__internal::__stable_sort_with_projection<oneapi::dpl::__internal::__device_backend_tag, oneapi::dpl::execution::device_policy<> &, int *, CompareWrapper<UserComparator, int>, oneapi::dpl::identity>' requested here
 1249 |     __stable_sort_with_projection(__tag, ::std::forward<_ExecutionPolicy>(__exec), __first, __last, __comp,
      |     ^
/home/darndt/onedpl-install/include/oneapi/dpl/pstl/glue_algorithm_impl.h:667:30: note: in instantiation of function template specialization 'oneapi::dpl::__internal::__pattern_sort<oneapi::dpl::__internal::__device_backend_tag, oneapi::dpl::execution::device_policy<> &, int *, CompareWrapper<UserComparator, int>>' requested here
  667 |     oneapi::dpl::__internal::__pattern_sort(__dispatch_tag, ::std::forward<_ExecutionPolicy>(__exec), __first, __last,
      |                              ^
test_Onedpl.cpp:69:16: note: in instantiation of function template specialization 'oneapi::dpl::sort<oneapi::dpl::execution::device_policy<> &, int *, CompareWrapper<UserComparator, int>>' requested here
   69 |   oneapi::dpl::sort(policy, values_ptr, values_ptr + n, comparator);
      |                ^
In file included from test_Onedpl.cpp:1:
In file included from /soft/preview/pe/24.180.0-RC5/CNDA/oneapi/compiler/eng-20240629/bin/compiler/../../include/sycl/sycl.hpp:25:
In file included from /soft/preview/pe/24.180.0-RC5/CNDA/oneapi/compiler/eng-20240629/bin/compiler/../../include/sycl/detail/core.hpp:21:
In file included from /soft/preview/pe/24.180.0-RC5/CNDA/oneapi/compiler/eng-20240629/bin/compiler/../../include/sycl/accessor.hpp:15:
In file included from /soft/preview/pe/24.180.0-RC5/CNDA/oneapi/compiler/eng-20240629/bin/compiler/../../include/sycl/buffer.hpp:19:
/soft/preview/pe/24.180.0-RC5/CNDA/oneapi/compiler/eng-20240629/bin/compiler/../../include/sycl/detail/is_device_copyable.hpp:109:17: error: static assertion failed due to requirement 'is_device_copyable_v<oneapi::dpl::__par_backend_hetero::__leaf_sorter<oneapi::dpl::__ranges::guard_view<int *>, oneapi::dpl::__internal::__compare<CompareWrapper<UserComparator, int>, oneapi::dpl::identity>>> || detail::IsDeprecatedDeviceCopyable<oneapi::dpl::__par_backend_hetero::__leaf_sorter<oneapi::dpl::__ranges::guard_view<int *>, oneapi::dpl::__internal::__compare<CompareWrapper<UserComparator, int>, oneapi::dpl::identity>>, void>::value': The specified type is not device copyable
  109 |   static_assert(is_device_copyable_v<FieldT> ||
      |                 ^~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~
  110 |                     detail::IsDeprecatedDeviceCopyable<FieldT>::value,
      |                     ~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~~
/soft/preview/pe/24.180.0-RC5/CNDA/oneapi/compiler/eng-20240629/bin/compiler/../../include/sycl/detail/is_device_copyable.hpp:107:7: note: in instantiation of template class 'sycl::detail::CheckFieldsAreDeviceCopyable<(lambda at /home/darndt/onedpl-install/include/oneapi/dpl/pstl/hetero/../hetero/dpcpp/parallel_backend_sycl_merge_sort.h:234:29), 1>' requested here
  107 |     : CheckFieldsAreDeviceCopyable<T, NumFieldsToCheck - 1> {
      |       ^
/soft/preview/pe/24.180.0-RC5/CNDA/oneapi/compiler/eng-20240629/bin/compiler/../../include/sycl/detail/is_device_copyable.hpp:142:7: note: in instantiation of template class 'sycl::detail::CheckFieldsAreDeviceCopyable<(lambda at /home/darndt/onedpl-install/include/oneapi/dpl/pstl/hetero/../hetero/dpcpp/parallel_backend_sycl_merge_sort.h:234:29), 2>' requested here
  142 |     : CheckFieldsAreDeviceCopyable<FuncT, __builtin_num_fields(FuncT)>,
      |       ^
/soft/preview/pe/24.180.0-RC5/CNDA/oneapi/compiler/eng-20240629/bin/compiler/../../include/sycl/handler.hpp:1765:5: note: in instantiation of template class 'sycl::detail::CheckDeviceCopyable<(lambda at /home/darndt/onedpl-install/include/oneapi/dpl/pstl/hetero/../hetero/dpcpp/parallel_backend_sycl_merge_sort.h:234:29)>' requested here
 1765 |     detail::CheckDeviceCopyable<KernelType>();
      |     ^
/soft/preview/pe/24.180.0-RC5/CNDA/oneapi/compiler/eng-20240629/bin/compiler/../../include/sycl/handler.hpp:1807:5: note: in instantiation of function template specialization 'sycl::handler::unpack<(lambda at /home/darndt/onedpl-install/include/oneapi/dpl/pstl/hetero/../hetero/dpcpp/parallel_backend_sycl_merge_sort.h:234:29), (lambda at /home/darndt/onedpl-install/include/oneapi/dpl/pstl/hetero/../hetero/dpcpp/parallel_backend_sycl_merge_sort.h:234:29), sycl::ext::oneapi::experimental::properties<std::tuple<>>, false, (lambda at /soft/preview/pe/24.180.0-RC5/CNDA/oneapi/compiler/eng-20240629/bin/compiler/../../include/sycl/handler.hpp:1810:21)>' requested here
 1807 |     unpack<KernelName, KernelType, PropertiesT,
      |     ^
/soft/preview/pe/24.180.0-RC5/CNDA/oneapi/compiler/eng-20240629/bin/compiler/../../include/sycl/handler.hpp:1480:5: note: in instantiation of function template specialization 'sycl::handler::kernel_parallel_for_wrapper<(lambda at /home/darndt/onedpl-install/include/oneapi/dpl/pstl/hetero/../hetero/dpcpp/parallel_backend_sycl_merge_sort.h:234:29), sycl::nd_item<>, (lambda at /home/darndt/onedpl-install/include/oneapi/dpl/pstl/hetero/../hetero/dpcpp/parallel_backend_sycl_merge_sort.h:234:29), sycl::ext::oneapi::experimental::properties<std::tuple<>>>' requested here
 1480 |     kernel_parallel_for_wrapper<NameT, TransformedArgType, KernelType,
      |     ^
/soft/preview/pe/24.180.0-RC5/CNDA/oneapi/compiler/eng-20240629/bin/compiler/../../include/sycl/handler.hpp:2506:5: note: (skipping 1 context in backtrace; use -ftemplate-backtrace-limit=0 to see all)
 2506 |     parallel_for_impl<KernelName>(Range, Properties, std::move(KernelFunc));
      |     ^
/soft/preview/pe/24.180.0-RC5/CNDA/oneapi/compiler/eng-20240629/bin/compiler/../../include/sycl/handler.hpp:2591:5: note: in instantiation of function template specialization 'sycl::handler::parallel_for<sycl::detail::auto_name, (lambda at /home/darndt/onedpl-install/include/oneapi/dpl/pstl/hetero/../hetero/dpcpp/parallel_backend_sycl_merge_sort.h:234:29), sycl::ext::oneapi::experimental::properties<std::tuple<>>, 1>' requested here
 2591 |     parallel_for<KernelName>(Range,
      |     ^
/home/darndt/onedpl-install/include/oneapi/dpl/pstl/hetero/../hetero/dpcpp/parallel_backend_sycl_merge_sort.h:233:19: note: in instantiation of function template specialization 'sycl::handler::parallel_for<sycl::detail::auto_name, 1, (lambda at /home/darndt/onedpl-install/include/oneapi/dpl/pstl/hetero/../hetero/dpcpp/parallel_backend_sycl_merge_sort.h:234:29)>' requested here
  233 |             __cgh.parallel_for<_LeafSortName...>(
      |                   ^
/home/darndt/onedpl-install/include/oneapi/dpl/pstl/hetero/algorithm_impl_hetero.h:1249:5: note: in instantiation of function template specialization 'oneapi::dpl::__internal::__stable_sort_with_projection<oneapi::dpl::__internal::__device_backend_tag, oneapi::dpl::execution::device_policy<> &, int *, CompareWrapper<UserComparator, int>, oneapi::dpl::identity>' requested here
 1249 |     __stable_sort_with_projection(__tag, ::std::forward<_ExecutionPolicy>(__exec), __first, __last, __comp,
      |     ^
/home/darndt/onedpl-install/include/oneapi/dpl/pstl/glue_algorithm_impl.h:667:30: note: in instantiation of function template specialization 'oneapi::dpl::__internal::__pattern_sort<oneapi::dpl::__internal::__device_backend_tag, oneapi::dpl::execution::device_policy<> &, int *, CompareWrapper<UserComparator, int>>' requested here
  667 |     oneapi::dpl::__internal::__pattern_sort(__dispatch_tag, ::std::forward<_ExecutionPolicy>(__exec), __first, __last,
      |                              ^
test_Onedpl.cpp:69:16: note: in instantiation of function template specialization 'oneapi::dpl::sort<oneapi::dpl::execution::device_policy<> &, int *, CompareWrapper<UserComparator, int>>' requested here
   69 |   oneapi::dpl::sort(policy, values_ptr, values_ptr + n, comparator);
      |                ^
2 errors generated.

indicating that comparators that aren't trivially-copyable but are made sycl::device_copyable explicitly aren't supported.

Supposedly, this issue was fixed with https://github.com/oneapi-src/oneDPL/pull/1621 but I'm still observing issues with https://github.com/oneapi-src/oneDPL/commits/oneDPL-2022.7.0-rc1/.

To Reproduce: see above

The following information might be useful:

• CMake command: doesn't apply
• oneDPL version: oneDPL-2022.7.0-rc1
• Compiler version: Intel(R) oneAPI DPC++/C++ Compiler 2025.0.0 (2025.x.0.20240629)
• OS: SUSE Linux Enterprise Server 15 SP4
• Device and driver version: Intel(R) Data Center GPU Max 1550 1.3 [1.3.28202]

Expected Behavior: The file should compile. All classes using comparators should be marked as sycl::device_copyable if the user provided classes are.

Additional Context: N/A

[dmitriy-sobolev]

@masterleinad, I've checked that the reproducer passes after applying #1932. I hope your main problem has also been fixed. As usual, feel free to reopen that issue or create a new one if the fix does not work for you.