cuda 推力::for_each 与推力::counting_iterator

Posted 2023-02-19

【中文标题】cuda 推力::for_each 与推力::counting_iterator

【英文标题】：cuda thrust::for_each with thrust::counting_iterator

【发布时间】：2017-02-14 03:34:24

【问题描述】：

我对 CUDA 和推力有点陌生。当提供counting_iterator 时，我似乎无法让thrust::for_each 算法工作。 这是我的简单函子：

struct print_Functor 
    print_Functor()
    __host__ __device__
    void operator()(int i)
    
        printf("index %d\n", i);
    
; 

现在，如果我使用预先填充了序列的宿主向量来调用它，它可以正常工作：

    thrust::host_vector<int> h_vec(10);
    thrust::sequence(h_vec.begin(),h_vec.end());
    thrust::for_each(h_vec.begin(),h_vec.end(), print_Functor());

但是，如果我尝试使用推力::counting_iterator 执行此操作，它会失败：

    thrust::counting_iterator<int> first(0);
    thrust::counting_iterator<int> last = first+10;
    for(thrust::counting_iterator<int> it=first;it!=last;it++)
        printf("Value %d\n", *it);
    printf("Launching for_each\n");
    thrust::for_each(first,last,print_Functor());

我得到的是 for 循环正确执行，但 for_each 失败并显示错误消息：

   after cudaFuncGetAttributes: unspecified launch failure

我试图通过使迭代器类型成为模板参数来做到这一点：

thrust::for_each<thrust::counting_iterator<int>>(first,last, print_Functor());

但同样的错误结果。

为了完整起见，我从 MATLAB mex 文件（64 位）中调用它。

我已经能够让其他推力算法与计数迭代器一起使用（例如，thrust::reduce 给出了正确的结果）。

作为一个新人，我可能正在做一些非常愚蠢的事情并且遗漏了一些明显的东西 - 任何人都可以帮忙吗？

感谢到目前为止的 cmets。到目前为止，我已经接受了 cmets。工作示例（在 Matlab 之外）正常工作并产生输出，但如果将其制成 mex 文件，它仍然无法工作 - 第一次根本不产生输出，第二次只产生与以前相同的错误消息（仅通过重新编译修复，当它返回无输出时）。

但是，即使在 DOS 下，它也不执行来自thrust::for_each 的函子也存在类似的问题。这是一个完整的例子：

#include <thrust/for_each.h>
#include <thrust/iterator/counting_iterator.h>

struct sum_Functor 
    int *sum;
    sum_Functor(int *s)sum = s;
    __host__ __device__
    void operator()(int i)
    
        *sum+=i;
        printf("In functor: i %d sum %d\n",i,*sum);
    

;

int main()

    thrust::counting_iterator<int> first(0);
    thrust::counting_iterator<int> last = first+10;
    int sum = 0;
    sum_Functor sf(&sum);
    printf("After constructor: value is %d\n", *(sf.sum));
    for(int i=0;i<5;i++)
        sf(i);
    

    printf("Initiating for_each call - current value %d\n", (*(sf.sum)));
    thrust::for_each(first,last,sf);

    cudaDeviceSynchronize();
    printf("After for_each: value is %d\n",*(sf.sum));

这是在 DOS 提示符下编译的：

nvcc -o pf pf.cu

产生的输出是：

After constructor: value is 0
In functor: i 0 sum 0
In functor: i 1 sum 1
In functor: i 2 sum 3
In functor: i 3 sum 6
In functor: i 4 sum 10
Initiating for_each call - current value 10
After for_each: value is 10

换句话说，仿函数的重载 operator() 在 for 循环中被正确调用，但从不被推力::for_each 算法调用。使用计数迭代器时，让 for_each 执行函子的唯一方法是省略成员变量。

（我应该补充一点，在使用纯 Matlab 多年后，我的 C++ 非常生锈，所以我可能会遗漏一些明显的东西......）

【问题讨论】：

您在哪个平台上？ linux？视窗？苹果？哪个CUDA版本？哪个推力版本？哪个编译器？请始终发布minimal reproducible example

here 是一个基于您所展示内容的工作（Linux 非 matlab）示例。它似乎对我来说工作正常。从 matlab mex 打印东西可能很棘手。

@m.s - 我在使用 CUDA 7.5 工具包的 Windows 7 Professional 64 位。不知道如何获得推力版本？随 CUDA 7.5 提供。

@RobertCrovella - 感谢您提供的示例。这在 DOS 下编译时可以正常工作，但在 Matlab 下不能。我得到的是第一次根本没有输出，第二次我得到完整的错误消息：MEX 文件中的意外标准异常。 What() is:function_attributes(): after cudaFuncGetAttributes: unspecified launch failure ..

thrust/version.h 中给出的推力版本表明我正在运行 1.8.2

【参考方案1】：

在您的 cmets 上，您说您希望您的代码在主机端执行。

错误代码“未指定的启动失败”，以及你的仿函数被定义为 host device 的事实让我觉得推力想在你的设备上执行。

你能添加一个执行策略来确定你的代码在哪里执行吗？

替换：

thrust::for_each(first,last,sf);

与

thrust::for_each(thrust::host, first,last,sf);

为了能够在 GPU 上运行，您的结果必须分配到设备内存上（通过 cudaMalloc），然后复制回主机。

---

#include <thrust/host_vector.h>
#include <thrust/sequence.h>
#include <thrust/for_each.h>
#include <thrust/iterator/counting_iterator.h>
#include <thrust/execution_policy.h>

struct sum_Functor 
    int *sum;
    sum_Functor(int *s)sum=s;
    __host__ __device__
    void operator()(int i)
    
        atomicAdd(sum, 1);
    
;

int main(int argc, char**argv)


    thrust::counting_iterator<int> first(0);
    thrust::counting_iterator<int> last = first+atoi(argv[1]);
    int *d_sum;
    int h_sum = 0;

    cudaMalloc(&d_sum,sizeof(int));
    cudaMemcpy(d_sum,&h_sum,sizeof(int),cudaMemcpyHostToDevice);

    thrust::for_each(thrust::device,first,last,sum_Functor(d_sum));

    cudaDeviceSynchronize();
    cudaMemcpy(&h_sum,d_sum,sizeof(int),cudaMemcpyDeviceToHost);
    printf("sum = %d\n", *h_sum);
    cudaFree(d_sum);

代码更新：要在您的设备上获得正确的结果，您必须使用原子操作。

【讨论】：

谢谢！是的，当您添加推力::主机参数时，现在确实会给出正确的输出。当然实际上我希望代码在设备上运行，但为了说明问题而对其进行了简化！看起来我在简化方面走得太远了。现在将通过分配设备内存来尝试。

我已经更新了你的代码，我不太喜欢一个 int 的 malloc，但最重要的是，你忘记了你的 int 的 AtomicOperation。

嗨，感谢@X3liF 的更新 - AtomicOperations 是我从交流中学到的新东西 - 作为 CUDA 的新手，我仍然需要了解一些领域。感谢所有的帮助。

欢迎您，不要忘记接受答案，这样其他人就会看到您的问题的解决方案。

不确定我必须做什么才能“接受”答案。我按下了答案旁边的“upvote”箭头（“答案很有用”），但它说我没有足够的经验值来注册投票。