推力：使用device_ptr时如何获取copy_if函数复制的元素个数

Posted 2023-03-23

【中文标题】推力：使用device_ptr时如何获取copy_if函数复制的元素个数

【英文标题】：Thrust: how to get the number of elements copied by the copy_if function when using device_ptr

【发布时间】：2013-11-03 07:31:46

【问题描述】：

我正在使用 Thrust 库的thrust::copy_if 函数和计数迭代器来获取数组中非零元素的索引。我还需要获取复制元素的数量。

我正在使用“counting_iterator.cu”示例中的代码，除了在我的应用程序中我需要重用预先分配的数组，所以我用thrust::device_ptr 包装它们，然后将它们传递给thrust::copy_if功能。这是代码：

using namespace thrust;

int output[5];
thrust::device_ptr<int> tp_output = device_pointer_cast(output);

float stencil[5];
stencil[0] = 0;
stencil[1] = 0;
stencil[2] = 1;
stencil[3] = 0;
stencil[4] = 1;
device_ptr<float> tp_stencil = device_pointer_cast(stencil);

device_vector<int>::iterator output_end = copy_if(make_counting_iterator<int>(0), 
     make_counting_iterator<int>(5), 
     tp_stencil, 
     tp_output, 
     _1 == 1);

int number_of_ones = output_end - tp_output;

如果我注释最后一行代码，函数会正确填充输出数组。但是，当我取消注释时，会出现以下编译错误：

1>C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v5.5\include\thrust/iterator/iterator_adaptor.h(223): error : no operator "-" 匹配这些操作数

1> 操作数类型为：int *const - constthrust::device_ptr

1> 在以下期间检测到： 1> 实例化“thrust::iterator_adaptor::difference_typethrust::iterator_adaptor::distance_to(const推进::iterator_adaptor &) const [with Derived=thrust::detail::normal_iterator>, Base=thrust::device_ptr, Value=推力::use_default, System=thrust::use_default, Traversal=thrust::use_default, Reference=thrust::use_default, Difference=thrust::use_default, OtherDerived=thrust::device_ptr, OtherIterator=int *, V=signed int, S=thrust::device_system_tag，T=thrust::random_access_traversal_tag，R=thrust::device_reference，D=ptrdiff_t]" 1> C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v5.5\include\thrust/iterator/iterator_facade.h(181): 这里 1> 实例化“Facade1::difference_typethrust::iterator_core_access::distance_from(const Facade1 &, const Facade2 &,thrust::detail::true_type) [with Facade1=thrust::detail::normal_iterator>, Facade2=thrust: :device_ptr]" 1> C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v5.5\include\thrust/iterator/iterator_facade.h(202): 这里 1> 实例化“thrust::detail::distance_from_result::typethrust::iterator_core_access::distance_from(const Facade1 &, const Facade2 &) [with Facade1=thrust::detail::normal_iterator>, Facade2=thrust::device_ptr ]" 1> C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v5.5\include\thrust/iterator/iterator_facade.h(506)：这里 1> 实例化“thrust::detail::distance_from_result,thrust::iterator_facade>::typethrust::operator-(constthrust::iterator_facade &, constthrust::iterator_facade &) [with Derived1=thrust::detail: :normal_iterator>, Value1=signed int, System1=thrust::device_system_tag, Traversal1=thrust::random_access_traversal_tag, Reference1=thrust::device_reference, Difference1=signed int, Derived2=thrust::device_ptr, Value2=signed int, System2=thrust ::device_system_tag, Traversal2=thrust::random_access_traversal_tag, Reference2=thrust::device_reference, Difference2=signed int]" 1> C:/ProgramData/NVIDIA Corporation/CUDA Samples/v5.5/7_CUDALibraries/nsgaIIparallelo_23ott/rank_cuda.cu(70)：这里

如果我使用thrust::device_vector 代替输出数组，一切正常：

using namespace thrust;

thrust::device_vector<int> output(5);

float stencil[5];
stencil[0] = 0;
stencil[1] = 0;
stencil[2] = 1;
stencil[3] = 0;
stencil[4] = 1;
device_ptr<float> tp_stencil = device_pointer_cast(stencil);

device_vector<int>::iterator output_end = copy_if(make_counting_iterator<int>(0), 
     make_counting_iterator<int>(5), 
     tp_stencil, 
     output.begin(), 
     _1 == 1);

int number_of_ones = output_end - output.begin();

您能提出任何解决此问题的方法吗？谢谢。

【参考方案1】：

尝试在您的 copy_if 调用中使用 device_ptr 而不是 device_vector::iterator：

thrust::device_ptr<int> output_end = copy_if(make_counting_iterator<int>(0),
 make_counting_iterator<int>(5),
 tp_stencil,
 tp_output,
 _1 == 1); 

【讨论】：

谢谢您，我已经尝试了您的建议，并且有效！实际上，我已经尝试过这种方式，但是我将减法的结果直接分配给数组的位置，并且由于某种奇怪的原因，它没有假定正确的值。相反，现在我尝试将减法的结果分配给一个临时整数变量，并稍后将其复制到数组中；这种方式效果很好。

我的 number_of_ones 总是 512 的倍数，即使它应该小于 10。我不知道为什么。