thrust/functional.h
File members: thrust/functional.h
/*
* Copyright 2008-2018 NVIDIA Corporation
*
* Licensed under the Apache License, Version 2.0 (the "License");
* you may not use this file except in compliance with the License.
* You may obtain a copy of the License at
*
* http://www.apache.org/licenses/LICENSE-2.0
*
* Unless required by applicable law or agreed to in writing, software
* distributed under the License is distributed on an "AS IS" BASIS,
* WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
* See the License for the specific language governing permissions and
* limitations under the License.
*/
#pragma once
#include <thrust/detail/config.h>
#if defined(_CCCL_IMPLICIT_SYSTEM_HEADER_GCC)
# pragma GCC system_header
#elif defined(_CCCL_IMPLICIT_SYSTEM_HEADER_CLANG)
# pragma clang system_header
#elif defined(_CCCL_IMPLICIT_SYSTEM_HEADER_MSVC)
# pragma system_header
#endif // no system header
#include <thrust/detail/functional/actor.h>
#include <cuda/functional>
#include <cuda/std/functional>
#include <functional>
THRUST_NAMESPACE_BEGIN
template <typename Operation>
struct THRUST_DEPRECATED unary_traits;
template <typename Operation>
struct THRUST_DEPRECATED binary_traits;
template <typename Argument, typename Result>
struct THRUST_DEPRECATED unary_function
{
using argument_type = Argument;
using result_type = Result;
}; // end unary_function
template <typename Argument1, typename Argument2, typename Result>
struct THRUST_DEPRECATED binary_function
{
using first_argument_type = Argument1;
using second_argument_type = Argument2;
using result_type = Result;
}; // end binary_function
#define THRUST_BINARY_FUNCTOR_VOID_SPECIALIZATION(func, impl) \
template <> \
struct func<void> \
{ \
using is_transparent = void; \
_CCCL_EXEC_CHECK_DISABLE \
template <typename T1, typename T2> \
_CCCL_HOST_DEVICE constexpr auto operator()(T1&& t1, T2&& t2) const noexcept(noexcept(impl)) -> decltype(impl) \
{ \
return impl; \
} \
}
template <typename T = void>
struct plus : public ::cuda::std::plus<T>
{
using first_argument_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
using second_argument_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
using result_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
}; // end plus
template <typename T = void>
struct minus : public ::cuda::std::minus<T>
{
using first_argument_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
using second_argument_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
using result_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
}; // end minus
template <typename T = void>
struct multiplies : public ::cuda::std::multiplies<T>
{
using first_argument_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
using second_argument_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
using result_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
}; // end multiplies
template <typename T = void>
struct divides : public ::cuda::std::divides<T>
{
using first_argument_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
using second_argument_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
using result_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
}; // end divides
template <typename T = void>
struct modulus : public ::cuda::std::modulus<T>
{
using first_argument_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
using second_argument_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
using result_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
}; // end modulus
template <typename T = void>
struct negate : ::cuda::std::negate<T>
{
using argument_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
using result_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
}; // end negate
template <typename T = void>
struct square
{
using argument_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
using result_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
_CCCL_EXEC_CHECK_DISABLE
_CCCL_HOST_DEVICE constexpr T operator()(const T& x) const
{
return x * x;
}
}; // end square
template <>
struct square<void>
{
using is_transparent = void;
_CCCL_EXEC_CHECK_DISABLE
template <typename T>
_CCCL_HOST_DEVICE constexpr T operator()(const T& x) const noexcept(noexcept(x * x))
{
return x * x;
}
};
template <typename T = void>
struct equal_to : public ::cuda::std::equal_to<T>
{
using first_argument_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
using second_argument_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
using result_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
}; // end equal_to
template <typename T = void>
struct not_equal_to : public ::cuda::std::not_equal_to<T>
{
using first_argument_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
using second_argument_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
using result_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
}; // end not_equal_to
template <typename T = void>
struct greater : public ::cuda::std::greater<T>
{
using first_argument_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
using second_argument_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
using result_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
}; // end greater
template <typename T = void>
struct less : public ::cuda::std::less<T>
{
using first_argument_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
using second_argument_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
using result_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
}; // end less
template <typename T = void>
struct greater_equal : public ::cuda::std::greater_equal<T>
{
using first_argument_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
using second_argument_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
using result_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
}; // end greater_equal
template <typename T = void>
struct less_equal : public ::cuda::std::less_equal<T>
{
using first_argument_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
using second_argument_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
using result_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
}; // end less_equal
template <typename T = void>
struct logical_and : public ::cuda::std::logical_and<T>
{
using first_argument_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
using second_argument_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
using result_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
}; // end logical_and
template <typename T = void>
struct logical_or : public ::cuda::std::logical_or<T>
{
using first_argument_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
using second_argument_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
using result_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
}; // end logical_or
template <typename T = void>
struct logical_not : public ::cuda::std::logical_not<T>
{
using first_argument_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
using second_argument_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
using result_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
}; // end logical_not
template <typename T = void>
struct bit_and : public ::cuda::std::bit_and<T>
{
using first_argument_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
using second_argument_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
using result_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
}; // end bit_and
template <typename T = void>
struct bit_or : public ::cuda::std::bit_or<T>
{
using first_argument_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
using second_argument_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
using result_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
}; // end bit_or
template <typename T = void>
struct bit_xor : public ::cuda::std::bit_xor<T>
{
using first_argument_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
using second_argument_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
using result_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
}; // end bit_xor
// TODO(bgruber): this version can also act as a functor casting to T making it not equivalent to ::cuda::std::identity
template <typename T = void>
struct identity
{
using argument_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
using result_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
_CCCL_EXEC_CHECK_DISABLE
_CCCL_HOST_DEVICE constexpr const T& operator()(const T& x) const
{
return x;
}
_CCCL_EXEC_CHECK_DISABLE
_CCCL_HOST_DEVICE constexpr T& operator()(T& x) const
{
return x;
}
// we cannot add an overload for `const T&&` because then calling e.g. `thrust::identity<int>{}(3.14);` is ambiguous
// on MSVC
_CCCL_EXEC_CHECK_DISABLE
_CCCL_HOST_DEVICE constexpr T&& operator()(T&& x) const
{
return _CUDA_VSTD::move(x);
}
};
template <>
struct identity<void> : ::cuda::std::__identity
{};
template <typename T = void>
struct maximum : ::cuda::maximum<T>
{
using first_argument_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
using second_argument_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
using result_type _LIBCUDACXX_DEPRECATED_IN_CXX11 = T;
}; // end maximum
template <typename T = void>
struct minimum : ::cuda::minimum<T>
{
using first_argument_type _CCCL_ALIAS_ATTRIBUTE(THRUST_DEPRECATED) = T;
using second_argument_type _CCCL_ALIAS_ATTRIBUTE(THRUST_DEPRECATED) = T;
using result_type _CCCL_ALIAS_ATTRIBUTE(THRUST_DEPRECATED) = T;
}; // end minimum
template <typename T1 = void, typename T2 = void>
struct project1st
{
using first_argument_type _CCCL_ALIAS_ATTRIBUTE(THRUST_DEPRECATED) = T1;
using second_argument_type _CCCL_ALIAS_ATTRIBUTE(THRUST_DEPRECATED) = T2;
using result_type _CCCL_ALIAS_ATTRIBUTE(THRUST_DEPRECATED) = T1;
_CCCL_HOST_DEVICE constexpr const T1& operator()(const T1& lhs, const T2& /*rhs*/) const
{
return lhs;
}
}; // end project1st
template <>
struct project1st<void, void>
{
using is_transparent = void;
_CCCL_EXEC_CHECK_DISABLE
template <typename T1, typename T2>
_CCCL_HOST_DEVICE constexpr auto operator()(T1&& t1, T2&&) const
noexcept(noexcept(THRUST_FWD(t1))) -> decltype(THRUST_FWD(t1))
{
return THRUST_FWD(t1);
}
};
template <typename T1 = void, typename T2 = void>
struct project2nd
{
using first_argument_type _CCCL_ALIAS_ATTRIBUTE(THRUST_DEPRECATED) = T1;
using second_argument_type _CCCL_ALIAS_ATTRIBUTE(THRUST_DEPRECATED) = T2;
using result_type _CCCL_ALIAS_ATTRIBUTE(THRUST_DEPRECATED) = T2;
_CCCL_HOST_DEVICE constexpr const T2& operator()(const T1& /*lhs*/, const T2& rhs) const
{
return rhs;
}
}; // end project2nd
template <>
struct project2nd<void, void>
{
using is_transparent = void;
_CCCL_EXEC_CHECK_DISABLE
template <typename T1, typename T2>
_CCCL_HOST_DEVICE constexpr auto operator()(T1&&, T2&& t2) const
noexcept(noexcept(THRUST_FWD(t2))) -> decltype(THRUST_FWD(t2))
{
return THRUST_FWD(t2);
}
};
// odds and ends
template <typename Predicate>
struct THRUST_DEPRECATED unary_negate
{
using argument_type = typename Predicate::argument_type;
using result_type = bool;
_CCCL_HOST_DEVICE explicit unary_negate(Predicate p)
: pred(p)
{}
_CCCL_EXEC_CHECK_DISABLE
_CCCL_HOST_DEVICE bool operator()(const typename Predicate::argument_type& x)
{
return !pred(x);
}
Predicate pred;
}; // end unary_negate
_CCCL_SUPPRESS_DEPRECATED_PUSH
template <typename Predicate>
_CCCL_HOST_DEVICE
THRUST_DEPRECATED_BECAUSE("Use thrust::not_fn instead") unary_negate<Predicate> not1(const Predicate& pred);
_CCCL_SUPPRESS_DEPRECATED_POP
template <typename Predicate>
struct THRUST_DEPRECATED binary_negate
{
using first_argument_type = typename Predicate::first_argument_type;
using second_argument_type = typename Predicate::second_argument_type;
using result_type = bool;
_CCCL_HOST_DEVICE explicit binary_negate(Predicate p)
: pred(p)
{}
_CCCL_EXEC_CHECK_DISABLE
_CCCL_HOST_DEVICE bool operator()(const first_argument_type& x, const second_argument_type& y)
{
return !pred(x, y);
}
Predicate pred;
}; // end binary_negate
_CCCL_SUPPRESS_DEPRECATED_PUSH
template <typename BinaryPredicate>
_CCCL_HOST_DEVICE THRUST_DEPRECATED_BECAUSE("Use thrust::not_fn instead")
binary_negate<BinaryPredicate> not2(const BinaryPredicate& pred);
_CCCL_SUPPRESS_DEPRECATED_POP
namespace detail
{
template <typename F>
struct not_fun_t
{
F f;
template <typename... Ts>
_CCCL_HOST_DEVICE auto
operator()(Ts&&... args) noexcept(noexcept(!f(std::forward<Ts>(args)...))) -> decltype(!f(std::forward<Ts>(args)...))
{
return !f(std::forward<Ts>(args)...);
}
template <typename... Ts>
_CCCL_HOST_DEVICE auto operator()(Ts&&... args) const
noexcept(noexcept(!f(std::forward<Ts>(args)...))) -> decltype(!f(std::forward<Ts>(args)...))
{
return !f(std::forward<Ts>(args)...);
}
};
} // namespace detail
// TODO(bgruber): alias to ::cuda::std::not_fn in C++17
template <class F>
_CCCL_HOST_DEVICE auto not_fn(F&& f) -> detail::not_fun_t<::cuda::std::decay_t<F>>
{
return detail::not_fun_t<::cuda::std::decay_t<F>>{std::forward<F>(f)};
}
namespace placeholders
{
THRUST_INLINE_CONSTANT thrust::detail::functional::placeholder<0>::type _1;
THRUST_INLINE_CONSTANT thrust::detail::functional::placeholder<1>::type _2;
THRUST_INLINE_CONSTANT thrust::detail::functional::placeholder<2>::type _3;
THRUST_INLINE_CONSTANT thrust::detail::functional::placeholder<3>::type _4;
THRUST_INLINE_CONSTANT thrust::detail::functional::placeholder<4>::type _5;
THRUST_INLINE_CONSTANT thrust::detail::functional::placeholder<5>::type _6;
THRUST_INLINE_CONSTANT thrust::detail::functional::placeholder<6>::type _7;
THRUST_INLINE_CONSTANT thrust::detail::functional::placeholder<7>::type _8;
THRUST_INLINE_CONSTANT thrust::detail::functional::placeholder<8>::type _9;
THRUST_INLINE_CONSTANT thrust::detail::functional::placeholder<9>::type _10;
} // namespace placeholders
#undef THRUST_BINARY_FUNCTOR_VOID_SPECIALIZATION
THRUST_NAMESPACE_END
#ifndef _CCCL_DOXYGEN_INVOKED // Do not document
_LIBCUDACXX_BEGIN_NAMESPACE_CUDA
_LIBCUDACXX_MARK_CAN_COPY_ARGUMENTS(THRUST_NS_QUALIFIER::plus);
_LIBCUDACXX_MARK_CAN_COPY_ARGUMENTS(THRUST_NS_QUALIFIER::minus);
_LIBCUDACXX_MARK_CAN_COPY_ARGUMENTS(THRUST_NS_QUALIFIER::multiplies);
_LIBCUDACXX_MARK_CAN_COPY_ARGUMENTS(THRUST_NS_QUALIFIER::divides);
_LIBCUDACXX_MARK_CAN_COPY_ARGUMENTS(THRUST_NS_QUALIFIER::modulus);
_LIBCUDACXX_MARK_CAN_COPY_ARGUMENTS(THRUST_NS_QUALIFIER::negate);
_LIBCUDACXX_MARK_CAN_COPY_ARGUMENTS(THRUST_NS_QUALIFIER::bit_and);
//_LIBCUDACXX_MARK_CAN_COPY_ARGUMENTS(THRUST_NS_QUALIFIER::bit_not); // does not exist?
_LIBCUDACXX_MARK_CAN_COPY_ARGUMENTS(THRUST_NS_QUALIFIER::bit_or);
_LIBCUDACXX_MARK_CAN_COPY_ARGUMENTS(THRUST_NS_QUALIFIER::bit_xor);
_LIBCUDACXX_MARK_CAN_COPY_ARGUMENTS(THRUST_NS_QUALIFIER::equal_to);
_LIBCUDACXX_MARK_CAN_COPY_ARGUMENTS(THRUST_NS_QUALIFIER::not_equal_to);
_LIBCUDACXX_MARK_CAN_COPY_ARGUMENTS(THRUST_NS_QUALIFIER::less);
_LIBCUDACXX_MARK_CAN_COPY_ARGUMENTS(THRUST_NS_QUALIFIER::less_equal);
_LIBCUDACXX_MARK_CAN_COPY_ARGUMENTS(THRUST_NS_QUALIFIER::greater_equal);
_LIBCUDACXX_MARK_CAN_COPY_ARGUMENTS(THRUST_NS_QUALIFIER::greater);
_LIBCUDACXX_MARK_CAN_COPY_ARGUMENTS(THRUST_NS_QUALIFIER::logical_and);
_LIBCUDACXX_MARK_CAN_COPY_ARGUMENTS(THRUST_NS_QUALIFIER::logical_not);
_LIBCUDACXX_MARK_CAN_COPY_ARGUMENTS(THRUST_NS_QUALIFIER::logical_or);
_LIBCUDACXX_END_NAMESPACE_CUDA
#endif // _CCCL_DOXYGEN_INVOKED
#include <thrust/detail/functional.inl>
#include <thrust/detail/functional/operators.h>
#include <thrust/detail/type_traits/is_commutative.h>