Spaces:
Runtime error
Runtime error
File size: 3,309 Bytes
be11144 |
1 2 3 4 5 6 7 8 9 10 11 12 13 14 15 16 17 18 19 20 21 22 23 24 25 26 27 28 29 30 31 32 33 34 35 36 37 38 39 40 41 42 43 44 45 46 47 48 49 50 51 52 53 54 55 56 57 58 59 60 61 62 63 64 65 66 67 68 69 70 71 72 73 74 75 76 77 78 79 80 81 82 83 84 85 86 87 88 89 90 91 92 93 94 95 96 97 98 99 100 101 102 103 104 105 106 107 108 109 110 111 112 113 114 115 116 117 118 119 120 121 122 123 124 125 126 127 128 |
#include <unittest/unittest.h>
#include <thrust/system/cuda/memory.h>
#include <thrust/system/cpp/memory.h>
#include <thrust/memory.h>
#include <thrust/execution_policy.h>
#include <thrust/logical.h>
template<typename T1, typename T2>
bool are_same_type(const T1 &, const T2 &)
{
return false;
}
template<typename T>
bool are_same_type(const T &, const T &)
{
return true;
}
void TestSelectSystemCudaToCpp()
{
using thrust::system::detail::generic::select_system;
thrust::cuda::tag cuda_tag;
thrust::cpp::tag cpp_tag;
thrust::cuda_cub::cross_system<thrust::cuda::tag,thrust::cpp::tag> cuda_to_cpp(cuda_tag, cpp_tag);
// select_system(cuda::tag, thrust::host_system_tag) should return cuda_to_cpp
bool is_cuda_to_cpp = are_same_type(cuda_to_cpp, select_system(cuda_tag, cpp_tag));
ASSERT_EQUAL(true, is_cuda_to_cpp);
}
DECLARE_UNITTEST(TestSelectSystemCudaToCpp);
template<typename Iterator>
__global__ void get_temporary_buffer_kernel(size_t n, Iterator result)
{
*result = thrust::get_temporary_buffer<int>(thrust::seq, n);
}
template<typename Pointer>
__global__ void return_temporary_buffer_kernel(Pointer ptr, std::ptrdiff_t n)
{
thrust::return_temporary_buffer(thrust::seq, ptr, n);
}
void TestGetTemporaryBufferDeviceSeq()
{
const std::ptrdiff_t n = 9001;
typedef thrust::pointer<int, thrust::detail::seq_t> pointer;
typedef thrust::pair<pointer, std::ptrdiff_t> ptr_and_sz_type;
thrust::device_vector<ptr_and_sz_type> d_result(1);
get_temporary_buffer_kernel<<<1,1>>>(n, d_result.begin());
cudaError_t const err = cudaDeviceSynchronize();
ASSERT_EQUAL(cudaSuccess, err);
ptr_and_sz_type ptr_and_sz = d_result[0];
if(ptr_and_sz.second > 0)
{
ASSERT_EQUAL(ptr_and_sz.second, n);
const int ref_val = 13;
thrust::device_vector<int> ref(n, ref_val);
thrust::fill_n(thrust::device, ptr_and_sz.first, n, ref_val);
ASSERT_EQUAL(true, thrust::all_of(thrust::device, ptr_and_sz.first, ptr_and_sz.first + n, thrust::placeholders::_1 == ref_val));
return_temporary_buffer_kernel<<<1,1>>>(ptr_and_sz.first, ptr_and_sz.second);
cudaError_t const err = cudaDeviceSynchronize();
ASSERT_EQUAL(cudaSuccess, err);
}
}
DECLARE_UNITTEST(TestGetTemporaryBufferDeviceSeq);
template<typename Iterator>
__global__ void malloc_kernel(size_t n, Iterator result)
{
*result = static_cast<int*>(thrust::malloc(thrust::seq, sizeof(int) * n).get());
}
template<typename Pointer>
__global__ void free_kernel(Pointer ptr)
{
thrust::free(thrust::seq, ptr);
}
void TestMallocDeviceSeq()
{
const std::ptrdiff_t n = 9001;
typedef thrust::pointer<int, thrust::detail::seq_t> pointer;
thrust::device_vector<pointer> d_result(1);
malloc_kernel<<<1,1>>>(n, d_result.begin());
cudaError_t const err = cudaDeviceSynchronize();
ASSERT_EQUAL(cudaSuccess, err);
pointer ptr = d_result[0];
if(ptr.get() != 0)
{
const int ref_val = 13;
thrust::device_vector<int> ref(n, ref_val);
thrust::fill_n(thrust::device, ptr, n, ref_val);
ASSERT_EQUAL(true, thrust::all_of(thrust::device, ptr, ptr + n, thrust::placeholders::_1 == ref_val));
free_kernel<<<1,1>>>(ptr);
cudaError_t const err = cudaDeviceSynchronize();
ASSERT_EQUAL(cudaSuccess, err);
}
}
DECLARE_UNITTEST(TestMallocDeviceSeq);
|