在Memcpy上未指定的启动失败

Question

在Cuda中运行我的程序时遇到“未指定的启动失败”。 我检查了错误。

该程序是一个微分方程的解算器。它迭代TOTAL_ITER次。 ROOM_X和ROOM_Y是矩阵的宽度和高度。

这里是头，它的名字是“独家：

#define ITER_BETWEEN_SAVES 10000 
#define TOTAL_ITER 10000 
#define ROOM_X 2048 
#define ROOM_Y 2048 
#define SOURCE_DIM_X 200 
#define SOURCE_DIM_Y 1000 
#define ALPHA 1.11e-4 
#define DELTA_T 10 
#define H 0.1 
#include <stdio.h> 

void Matrix(float* M); 
void SolverCPU(float* M1, float* M2); 
__global__ void SolverGPU(float* M1, float* M2); 

这是在内核和功能，填补矩阵：

#include "solver.h" 
#include<cuda.h> 

void Matrix(float* M) 
{ 
    for (int j = 0; j < SOURCE_DIM_Y; ++j) { 
    for (int i = 0; i < SOURCE_DIM_X; ++i) { 
    M[(i+(ROOM_X/2 - SOURCE_DIM_X/2)) + ROOM_X * (j+(ROOM_Y/2 - SOURCE_DIM_Y/2))] = 100; 
    } 
    } 
} 

    __global__ void SolverGPU(float* M1,float *M2) { 
    int i =threadIdx.x + blockIdx.x * blockDim.x; 
     int j = threadIdx.y + blockIdx.y * blockDim.y; 

     float M1_Index = M1[i + ROOM_X * j]; 
     float M1_IndexUp = M1[i+1 + ROOM_X * j]; 
     float M1_IndexDown =M1[i-1 + ROOM_X * j]; 
     float M1_IndexLeft = M1[i + ROOM_X * (j+1)]; 
     float M1_IndexRight = M1[i + ROOM_X *(j-1)]; 


     M2[i + ROOM_X * j] = M1_Index + (ALPHA * DELTA_T/(H*H)) * (M1_IndexUp + M1_IndexDown + M1_IndexLeft +M1_IndexRight - 4*M1_Index);  

} 

这里是主

int main(int argc, char* argv[]){ 

    float *M1_h, *M1_d,*M2_h, *M2_d; 
    int size = ROOM_X * ROOM_Y * sizeof(float); 
    cudaError_t err = cudaSuccess; 

    //Allocating Memories on Host 
    M1_h = (float *)malloc(size); 
    M2_h = (float *)malloc(size); 

    //Allocating Memories on Host 
    err=cudaMalloc((void**)&M1_d, size); 
    if (err != cudaSuccess) { 
     fprintf(stderr, "Failed to allocate array_d ... %s .\n", cudaGetErrorString(err)); 
     exit(EXIT_FAILURE); 
    } 

    err=cudaMalloc((void**)&M2_d, size);  
    if (err != cudaSuccess) { 
     fprintf(stderr, "Failed to allocate array_d ... %s .\n", cudaGetErrorString(err)); 
     exit(EXIT_FAILURE); 
    } 

    //Filling the Matrix 
    Matrix(M1_h); 


    //Copy on Device 

    err = cudaMemcpy(M1_d, M1_h, size, cudaMemcpyHostToDevice); 
    if(err !=0){ 
     printf("%s-%d\n",cudaGetErrorString(err),1); 
     getchar(); 
    } 

    err=cudaMemcpy(M2_d, M2_h, size, cudaMemcpyHostToDevice); 
    if(err !=0){ 
     printf("%s-%d",cudaGetErrorString(err),2); 
     getchar(); 
    } 

    dim3 dimGrid(64,64); 
    dim3 dimBlock(32,32); 


    //SolverGPU<< <threadsPerBlock, numBlocks >> >(M1_d,M2_d); 
    for(int i=0;i<TOTAL_ITER;i++) { 
    if (i%2==0) 
    SolverGPU<< <dimGrid,dimBlock >> >(M1_d,M2_d); 
    else 
    SolverGPU<< <dimGrid,dimBlock >> >(M2_d,M1_d); 
    } 

    err=cudaMemcpy(M1_h, M1_d, size, cudaMemcpyDeviceToHost); 
    if(err !=0){ 
     printf("%s-%d",cudaGetErrorString(err),3); 
     getchar(); 
    } 

    cudaFree(M1_d); 
    cudaFree(M2_d); 

    free(M1_h); 
    free(M2_h); 
    return 0; 

} 

编译时没有问题

当我检查我的错误时，“未指定的启动失败”出现在内核之后的memcpy中。

好的，我读过它通常是由于内核不能正常运行。但是我无法在内核中找到错误...我想这是错误很简单，但无法找到它。

Answer 1

当我编译并运行你的代码，我得到：

an illegal memory access was encountered-3 

打印出来。

您可能确实正在获取“未指定的启动失败”。确切的错误报告将取决于CUDA版本，GPU和平台。但我们可以继续前进。

任一消息都表示内核启动但遇到错误，因此未能成功完成。您可以使用调试器来调试内核执行问题，例如Linux上的cuda-gdb或Windows上的Nsight VSE。但是我们不需要退出调试器。

一个有用的工具是cuda-memcheck。如果我们用cuda-memcheck运行程序，我们会得到一些额外的输出，表明内核正在执行大小为4的无效全局读取。这意味着您正在进行超出界限的内存访问。如果我们重新编译您的代码并添加-lineinfo开关，然后重新运行cuda-memcheck的代码，我们可以获得更多的清晰度。现在，我们得到的输出将是这样的：

$ nvcc -arch=sm_20 -lineinfo -o t615 t615.cu 
$ cuda-memcheck ./t615 |more 
========= CUDA-MEMCHECK 
========= Invalid __global__ read of size 4 
=========  at 0x00000070 in /home/bob/misc/t615.cu:34:SolverGPU(float*, float*) 
=========  by thread (31,0,0) in block (3,0,0) 
=========  Address 0x4024fe1fc is out of bounds 
=========  Saved host backtrace up to driver entry point at kernel launch time 
=========  Host Frame:/usr/lib64/libcuda.so.1 (cuLaunchKernel + 0x2cd) [0x150a7d] 
=========  Host Frame:./t615 [0x11ef8] 
=========  Host Frame:./t615 [0x3b143] 
=========  Host Frame:./t615 [0x297d] 
=========  Host Frame:./t615 (__gxx_personality_v0 + 0x378) [0x26a0] 
=========  Host Frame:./t615 (__gxx_personality_v0 + 0x397) [0x26bf] 
=========  Host Frame:./t615 [0x2889] 
=========  Host Frame:/lib64/libc.so.6 (__libc_start_main + 0xf4) [0x1d994] 
=========  Host Frame:./t615 (__gxx_personality_v0 + 0x111) [0x2439] 
========= 
--More-- 

（并且还有更多的错误输出）

这意味着，你的内核遇到的第一个错误是大小为4的无效全局读（即例如，尝试读取数量的int或float数量的越界访问）。使用lineinfo信息，我们可以看到发生了这种情况：

=========  at 0x00000070 in /home/bob/misc/t615.cu:34:SolverGPU(float*, float*) 

即在文件中的第34行。此行正好是该行内核代码：

float M1_IndexRight = M1[i + ROOM_X *(j-1)]; 

我们可以进一步调试，可能使用在内核printf语句来发现问题的所在。但是，我们已经有了，我们索引出界外的线索，让我们检查索引：

i + ROOM_X *(j-1) 

是什么评价时i = 0和j = 0（即线程（0， 0）在你的2D线程数组中）？它评估为-2048（即 - ROOM_X），这是一个非法指数。试图从M1[-2048]读取将创建一个错误。

你的内核里有很多复杂的索引，所以我敢肯定还有其他错误。您可以使用类似的方法来追踪那些（可能使用printf来吐出计算出的索引，或者测试索引的有效性）。