在 GPU 上使用两个以上的列表生成笛卡尔积

编程问答 2022-05-29

问题描述

我想知道如何使用 CUDA 生成两个以上列表的笛卡尔积。

如何使此代码适用于三个或更多列表？

它适用于两个列表但不适用于三个列表，我试过 /,% 没有成功。

它是基本的。

#include <thrust/device_vector.h>
    #include <thrust/pair.h>
    #include <thrust/copy.h>
    #include <iterator>
    
    __global__ void cartesian_product(const int *a,size_t a_size,const int *b,size_t b_size,const int *c,size_t c_size)
    {
      unsigned int idx = blockIdx.x * blockDim.x + threadIdx.x;
    
      if(idx < a_size * b_size * c_size) 
      {
        unsigned int a_idx = idx / a_size;
        unsigned int b_idx = idx % a_size;
        
        // ? 
        unsigned int c_idx = idx % a_size;
    
    
        
        printf("a[a_idx] and b[b_idx] and c[c_idx] are: %d %d %d\n\n",a[a_idx],b[b_idx],c[c_idx]);
        //1 3 5,1 3 6,1 4 5,1 4 6,2 3 5,2 3 6,2 4 5,2 4 6  
        //0 0 0,0 0 1,0 1 0,0 1 1,1 0 0,1 0 1,1 1 0,1 1 1
      }
    }
    
    int main()
    {
      
      
      // host_vector is stored in host memory while device_vector livesin GPU device memory.
      // a has storage for 2 integers
      thrust::device_vector<int> a(2);
      
      // initialize individual elements
      a[0] = 1; 
      a[1] = 2; 
     
    
      // b has storage for 2 integers
      thrust::device_vector<int> b(2);
      
      // initialize individual elements
      b[0] = 3; 
      b[1] = 4; 
     
    
       // d has storage for 2 integers
      thrust::device_vector<int> c(2);
      
      // initialize individual elements
      c[0] = 5; 
      c[1] = 6;
      
       
      unsigned int block_size = 256;
      unsigned int num_blocks = (8 + (block_size - 1)) / block_size;
    
    
      // raw_pointer_cast creates a "raw" pointer from a pointer-like type,simply returning the wrapped pointer,should it exist.
      cartesian_product<<<num_blocks,block_size>>>(thrust::raw_pointer_cast(a.data()),a.size(),thrust::raw_pointer_cast(b.data()),b.size(),thrust::raw_pointer_cast(c.data()),c.size());

      
      
      return 0;
    }

如果我想要三个以上的列表，如何在内核和后续数组中获得正确的 c_idx？

解决方法

在我看来你想要“词法索引”：

idx == (a_idx * b_size + b_idx) * c_size + c_idx

所以你的索引是这样的：

c_idx = idx % c_size;
b_idx = (idx / c_size) % b_size;
a_idx = (idx / c_size) / b_size;

这很容易推广到更多维度。例如。在四个维度中你有

idx == ((a_idx * b_size + b_idx) * c_size + c_idx) * d_size + d_idx

那么：

d_idx = idx % d_size;
c_idx = (idx / d_size) % c_size;
b_idx = ((idx / d_size) / c_size) % b_size;
a_idx = ((idx / d_size) / c_size) / b_size;

在 C/C++ 编程中，人们喜欢使用它来计算表示多维数据集的一维动态数组的索引。在 CUDA 中，您通常不需要那么多，因为 CUDA 最多可以为您提供三维 threadIdx/blockIdx/etc.. 因此对于三个数组的笛卡尔积，您将不需要这种技术，但只能使用固有的 CUDA 功能。即使在三个以上，性能最好的解决方案也会从内核的三个维度中的两个获取两个索引，并在第三个维度上使用词法索引：

__global__ void cartesian_product_5d(const int *a,size_t a_size,const int *b,size_t b_size,const int *c,size_t c_size,const int *d,size_t d_size,const int *e,size_t e_size)
{
    int idx = blockIdx.x * blockDim.x + threadIdx.x;
    int d_idx = blockIdx.y * blockDim.y + threadIdx.y;
    int e_idx = blockIdx.z * blockDim.z + threadIdx.z;
    /* idx == (c_idx * b_size + b_idx) * a_size + a_idx */
    int a_idx = idx % a_size;
    int b_idx = (idx / a_size) % b_size;
    int c_idx = (idx / a_size) / b_size;

    /* ... */
}
 
int main()
{
    /* ... */
    dim3 threadsPerBlock(8,8,8);
    dim3 numBlocks((a_size + b_size + c_size + threadsPerBlock.x - 1) /
                   threadsPerBlock.x,(d_size + threadsPerBlock.y - 1) / threadsPerBlock.y,(e_size + threadsPerBlock.z - 1) / threadsPerBlock.z);
    cartesian_product_5d<<<numBlocks,threadsPerBlock>>>(/* ... */);
    /* ... */
}
cartesiancudagpunumerical-methods