在二维数组上比较 Matlab 与 CUDA 的相关性和减少

Question

我正在尝试比较使用 FFT 与使用窗口方法的互相关。

我的 Matlab 代码是：

isize = 20;
n = 7;
for i = 1:n %%7x7 xcorr
  for j = 1:n
    xcout(i,j) = sum(sum(ffcorr1 .* ref(i:i+isize-1,j:j+isize-1))); %%ref is 676 element array and ffcorr1 is a 400 element array
  end
end

类似的CUDA内核：

__global__ void xc_corr(double* in_im, double* ref_im, int pix3, int isize, int n, double* out1, double* temp1, double* sum_temp1)
{

    int p = blockIdx.x * blockDim.x + threadIdx.x;
    int q = 0;
    int i = 0;
    int j = 0;
    int summ = 0;

    for(i = 0; i < n; ++i)
    {
        for(j = 0; j < n; ++j)
        {
            summ  = 0; //force update
            for(p = 0; p < pix1; ++p)
            {
                for(q = 0; q < pix1; ++q)
                {
                    temp1[((i*n+j)*pix1*pix1)+p*pix1+q] = in_im[p*pix1+q] * ref_im[(p+i)*pix1+(q+j)];               
                    sum_temp1[((i*n+j)*pix1*pix1)+p*pix1+q] += temp1[((i*n+j)*pix1*pix1)+p*pix1+q];
                    out1[i*n+j] = sum_temp1[((i*n+j)*pix1*pix1)+p*pix1+q];
                }
            }       
        }
    }

我在内核中将其称为

int blocksize = 64; //multiple of 32
int nblocks = (pix3+blocksize-1)/blocksize; //round to max pix3 = 400
xc_corr <<< nblocks,blocksize >>> (ffcorr1, ref_d, pix3, isize, npix, xcout, xc_partial);
cudaThreadSynchronize();

不知何故，当我对输出文件进行比较时，我看到 CUDA 内核只计算前 400 个元素。

编写这个内核的正确方法是什么？？

另外，在我的内核中声明 i,j 有什么区别？？

int i = blockIdx.x * blockDim.y + threadIdx.x * threadIdx.y;
int j = blockIdx.y * blockDim.x + threadIdx.x * threadIdx.y;

Answer 1

int blocksize = 64; //multiple of 32
int nblocks = (pix3+blocksize-1)/blocksize; //round to max pix3 = 400
xc_corr <<< nblocks,blocksize >>> (ffcorr1, ref_d, pix3, isize, npix, xcout, xc_partial);

意味着每个块启动 64 个线程，线程块的数量比处理 pix3 元素所需的数量多 1 个。如果 pix3 确实是 400，那么您正在处理 400 个元素，因为您将启动 7 个线程块，每个线程块执行 64 个点，其中 48 个不执行任何操作。

我不太确定这里有什么问题。

还有，

int i = blockIdx.x * blockDim.y + threadIdx.x * threadIdx.y;
int j = blockIdx.y * blockDim.x + threadIdx.x * threadIdx.y;

blocksize 和 nblocks 实际上被转换为 dim3 向量，因此它们具有 (x,y,z) 值。如果你用 > 调用内核，它会被翻译成

dim3 blocksize(64,1,1);
dim3 nblocks(7,1,1);
kernel<<blocksize,nblocks>>();

所以对于每个内核调用，blockIdx 有 3 个组件，线程 id x、y 和 z，对应于您所在的线程的 3d 网格。在您的情况下，因为您只有一个 x 组件，blockIdx。 y 和 threadIdx.y 无论如何都将是 1。所以本质上，它们是无用的。

老实说，您似乎应该重新阅读用户手册中的 CUDA 基础知识，因为您似乎缺少很多基础知识。在这里解释它是不经济的since it's all written down in a nice documentation you can get here。如果您只想通过 cuda 获得更快的 FFT，您可以下载并安装在 Nvidia 的 CUDA 专区上的许多库，如果您不关心学习 CUDA，这些库将为您完成。

祝你好运。

PS。你不需要在每个内核之后调用 cudaThreadSynchronize ;)