         cudaStream_t stream_ = (cudaStream_t)stream;
         cupy_allocator alloc(memory);
 
-        device_ptr<size_t> dp_idx_first, dp_idx_last;
-        device_ptr<T> dp_data_first, dp_data_last;
-        device_ptr<size_t> dp_keys_first, dp_keys_last;
+        thrust::device_ptr<size_t> dp_idx_first, dp_idx_last;
+        thrust::device_ptr<T> dp_data_first, dp_data_last;
+        thrust::device_ptr<size_t> dp_keys_first, dp_keys_last;
 
         // Compute the total size of the data array.
         size = shape[0];

#if (__CUDACC_VER_MAJOR__ >11 || (__CUDACC_VER_MAJOR__ == 11 && __CUDACC_VER_MINOR__ >= 2) || HIP_VERSION >= 402)
THRUST_OPTIONAL_CPP11_CONSTEXPR_LESS
#endif
bool less< tuple<size_t, float> >::operator() (
    const tuple<size_t, float>& lhs, const tuple<size_t, float>& rhs) const {

    return _tuple_less<float>(lhs, rhs);
}

#include "cupy_thrust.h"


#if CUPY_USE_HIP
typedef hipStream_t cudaStream_t;
namespace cuda {
    using thrust::hip::par;
}
#else // #if CUPY_USE_HIP
namespace cuda {
    using thrust::cuda::par;
}
#endif // #if CUPY_USE_HIP


