Memcpy 上未指定的启动失败

Posted 2023-03-23

【中文标题】Memcpy 上未指定的启动失败

【英文标题】：Unspecified launch failure on Memcpy

【发布时间】：2015-02-01 08:22:49

【问题描述】：

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

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

这是标题，它的名字是“sole:

#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 AFTER the kernel 上。

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

【问题讨论】：

一个带有不可编译代码的“为什么这行不通”的问题完全是在浪费每个人的时间。投票结束。

感谢您的友好回复！ :) 的确，我忘了放一个函数来删除一些代码行。真的很抱歉。我编辑了我的帖子。我希望它现在可以编译。

例如，在不知道ROOM_X 是什么的情况下，如何编译它？ SO expects 和 MCVE。它应该是一个完整的代码。如果您想测试您是否正确编写了问题，请启动一个新的空项目，并将代码从您发布的问题中复制出来，而不添加任何内容或更改任何内容（毕竟，其他人将要做什么）做）。然后看看能不能编译。如果不能，则您的问题/MCVE 不完整。

我做到了。感谢您的解释 。我想我可以自己找到它，我想我有点累了。x）现在这段代码应该可以编译了。

【参考方案1】：

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

an illegal memory access was encountered-3

打印出来。

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

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

一个有用的工具是cuda-memcheck。 （在较新的 GPU，例如 cc7.0 或更新的 GPU 上，您应该使用 compute-sanitizer 而不是 cuda-memcheck，否则这里的过程是相同的。）如果我们使用 cuda-memcheck 运行您的程序，我们会得到一些额外的输出，表明内核正在执行大小为 4 的无效全局读取。这意味着您正在进行越界内存访问。如果我们重新编译您的代码并添加-lineinfo 开关（或者使用-G），然后使用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)），这会产生什么影响？它评估为 -2048（即 -ROOM_X），这是一个非法索引。尝试从 M1[-2048] 读取将产生错误。

你的内核中有很多复杂的索引，所以我很确定还有其他错误。您可以使用类似的方法来追踪它们（也许使用printf 吐出计算的索引，或者测试索引的有效性）。