在CUDA Thrust的transform操作符()函数中获取向量的索引

Question

在CUDA Thrust transform中，是否可以在operator()函数内获取传入的向量的索引？例如，我们有以下内容：

struct op{
    float operator()(const float& f){
        //do something like return the index
    }
};
vector<float> v(100);
thrust::transform(v.begin(),v.end(),v.begin(),op());

如何在operator()内获取向量的索引？基本上，我想在CUDA中轻松创建一个恒等矩阵。

- TK Omble

Answer 1

可能有许多方法来实现这个问题。其中一种方法是：

1. 使用thrust::sequence创建一个与数据向量长度相同的索引向量（或者直接使用一个counting_iterator）
2. 使用zip_iterator返回一个thrust::tuple，将数据向量和索引向量组合在一起，返回数据项加其索引的元组
3. 定义操作符op()来将特定的元组作为参数之一
4. 在操作符中，使用thrust::get<>从元组中检索需要的数据元素或索引

您可以在thrust快速入门指南中了解更多关于这些概念的信息。 编辑：回答下面的问题，这里有一个示例。虽然这实际上没有使用任何device_vector，但如果我们在GPU上执行此操作（使用device_vector），那么只会生成对thrust::transform的调用，即只会有1次在GPU上进行的“传递”。

（是的，thrust::sequence调用也会生成一个GPU内核，但我只是用它来为这个示例创建一些数据）。

#include <thrust/host_vector.h>
#include <thrust/iterator/zip_iterator.h>
#include <thrust/iterator/counting_iterator.h>
#include <thrust/sequence.h>
#include <thrust/copy.h>

#define N 30
#define SELECT 3

typedef thrust::tuple<int, int>            tpl2int;
typedef thrust::host_vector<int>::iterator intiter;
typedef thrust::counting_iterator<int>     countiter;
typedef thrust::tuple<intiter, countiter>  tpl2intiter;
typedef thrust::zip_iterator<tpl2intiter>  idxzip;



struct select_unary_op : public thrust::unary_function<tpl2int, int>
{
  __host__ __device__
  int operator()(const tpl2int& x) const
  {
    if ((x.get<1>() %SELECT) == 0)
      return x.get<0>();
    else return -1;
   }
};

int main() {

  thrust::host_vector<int> A(N);
  thrust::host_vector<int> result(N);
  thrust::sequence(A.begin(), A.end());
  thrust::counting_iterator<int> idxfirst(0);
  thrust::counting_iterator<int> idxlast = idxfirst +N;

  idxzip first = thrust::make_zip_iterator(thrust::make_tuple(A.begin(), idxfirst));
  idxzip  last = thrust::make_zip_iterator(thrust::make_tuple(A.end(), idxlast));
  select_unary_op my_unary_op;

  thrust::transform(first, last, result.begin(), my_unary_op);
  std::cout << "Results :" << std::endl;
  thrust::copy(result.begin(), result.end(), std::ostream_iterator<int>( std::cout, " "));
  std::cout << std::endl;


  return 0;

}

- Robert Crovella

---

如果我在GPU中硬编码，它会增加到2个或更多通行证吗？如果我在原始CUDA中进行硬编码，我可以在一次通行中完成...感谢链接... - TK Omble