CUDA - 多次调用内核

我在尝试编写CUDA程序时遇到困难。我有一个大约524k浮点值（1.0）的数组，我使用缩减技术来添加所有的值。如果我只想运行一次，问题就可以正常工作，但我真的想多次运行内核，以便最终总结超过10亿个值。CUDA - 多次调用内核

我以524k为单位做这件事的原因是当我在GPU上超过100万时，我总是得到零。这应该不会超过卡上的内存，但在这一点上总是失败。

无论如何，当我循环内核只有一次，一切工作正常。也就是说，没有循环是好的。当我用循环运行时，它返回零。我怀疑我会走出一些地方，但我无法弄清楚。这让我疯狂。

任何帮助表示赞赏，

感谢，

铝

下面是代码：

#include <stdio.h> 
#include <stdlib.h> 
#include "cutil.h" 

#define TILE_WIDTH  512 
#define WIDTH   524288 
//#define WIDTH   1048576 
#define MAX_WIDTH  524288 

#define BLOCKS   WIDTH/TILE_WIDTH 

__global__ void PartSum(float * V_d) 
{ 
    int tx = threadIdx.x; 
    int bx = blockIdx.x; 

    __shared__ float partialSum[TILE_WIDTH]; 

    for(int i = 0; i < WIDTH/TILE_WIDTH; ++i) 
    { 
     partialSum[tx] = V_d[bx * TILE_WIDTH + tx]; 
     __syncthreads(); 


     for(unsigned int stride = 1; stride < blockDim.x; stride *= 2) 
     { 
     __syncthreads(); 
     if(tx % (2 * stride) == 0) 
      partialSum[tx] += partialSum[tx + stride]; 
     } 
    } 

    if(tx % TILE_WIDTH == 0) 
     V_d[bx * TILE_WIDTH + tx] = partialSum[tx]; 
} 

int main(int argc, char * argv[]) 
{ 
    float * V_d; 
    float * V_h; 
    float * R_h; 
    float * Result; 
    float * ptr; 

    dim3 dimBlock(TILE_WIDTH,1,1); 
    dim3 dimGrid(BLOCKS,1,1); 

    // Allocate memory on Host 
    if((V_h = (float *)malloc(sizeof(float) * WIDTH)) == NULL) 
    { 
     printf("Error allocating memory on host\n"); 
     exit(-1); 
    } 

    if((R_h = (float *)malloc(sizeof(float) * MAX_WIDTH)) == NULL) 
    { 
     printf("Error allocating memory on host\n"); 
     exit(-1); 
    } 

    // If MAX_WIDTH is not a multiple of WIDTH, this won't work 
    if(WIDTH % MAX_WIDTH != 0) 
    { 
     printf("The width of the vector must be a multiple of the maximum width\n"); 
     exit(-3); 
    } 

    // Initialize memory on host with 1.0f 
    ptr = V_h; 
    for(long long i = 0; i < WIDTH; ++i) 
    { 
     *ptr = 1.0f; 
     ptr = &ptr[1]; 
    } 

    ptr = V_h; 

    // Allocate memory on device in global memory 
    cudaMalloc((void**) &V_d, MAX_WIDTH*(sizeof(float))); 
    float Pvalue = 0.0f; 
    for(int i = 0; i < WIDTH/MAX_WIDTH; ++i) 
    { 


    if((Result = (float *) malloc(sizeof(float) * WIDTH)) == NULL) 
    { 
     printf("Error allocating memory on host\n"); 
     exit(-4); 
    } 

    for(int j = 0; j < MAX_WIDTH; ++j) 
    { 
     Result[j] = *ptr; 
     ptr = &ptr[1]; 
    } 

     ptr = &V_h[i*MAX_WIDTH]; 
     // Copy portion of data to device 
     cudaMemcpy(V_d, Result, MAX_WIDTH*(sizeof(float)), cudaMemcpyHostToDevice); 

     // Execute Kernel 
     PartSum<<<dimGrid, dimBlock>>>(V_d); 

     // Copy data back down to host 
     cudaMemcpy(R_h, V_d, MAX_WIDTH*(sizeof(float)), cudaMemcpyDeviceToHost); 

     for(int i = 0; i < MAX_WIDTH; i += TILE_WIDTH) 
     { 
     Pvalue += R_h[i]; 
     } 
printf("Pvalue == %f\n", Pvalue); 

    free(Result); 


    } 

// printf("WIDTH == %d items\n", WIDTH); 
// printf("Value: %f\n", Pvalue); 

    cudaFree(V_d); 
    free(V_h); 
    free(R_h); 
    return(1); 
}

好吧，我想我已经缩小到与问题V_d在设备上。我怀疑我已经超出了阵列的范围了。如果我分配了实际需要的内存量的2倍，程序将以预期的结果完成。问题是，我无法弄清楚造成问题的原因。

铝

来源

2011-11-29 Al H.

你有使用cudaMemcpyDeviceToHost的特殊原因吗？既然你想要更多的内核迭代，你可以考虑使用'cudaMemcpyDeviceToDevice'来代替。 – karlphillip

我不相信你真的需要CUDA - 这将会是I/O的主导因素，因为你只需要每点添加一个操作 - 你也可以使用CPU。你有没有对这个实现的CPU实现进行基准测试？你认为CUDA的实现可能会快多少，因为它全部是数据移动，几乎没有计算？ –

这对我来说是一个学习实验。我意识到这并不高效。 –

我觉得我看到的第一个错误的位置：

if(tx % TILE_WIDTH == 0) 
     V_d[bx * TILE_WIDTH + tx] = partialSum[tx];

的范围TX是0-511，它从来没有达到512所以如果条件永远不会真正。如果（tx％（TILE_WIDTH-1）== 0），可以将其写为。

来源

2011-11-30 09:20:22 jwdmsd

啊！我没有看到！谢谢。 –

首先，感谢大家给了这个看看和任何帮助。

其次，我终于弄清楚我做错了什么。 BLOCKS应该被定义为MAX_WIDTH/TILE_WIDTH，而不是WIDTH/TILE_WIDTH。我的部分愚蠢愚蠢的错误。

再次感谢。

来源

2011-11-30 21:03:29

CUDA - 多次调用内核

回答

相关问题