Allow non-trivially-copyable comparators for SYCL #6939
Comments
|
Are you able to reproduce with a small program that does not use Kokkos and call |
#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);
} |
Sign up for free
to join this conversation on GitHub.
Already have an account?
Sign in to comment
Related to kokkos/kokkos-core-wiki#504 and #6801 (comment).
Even with changes like
I struggle to convince the SYCL compiler to accept custom (non-trivially-copyable) comparators
The text was updated successfully, but these errors were encountered: