cuda - blockDim.xにアクセスできない？

-1

私は、2D画像を処理するためのプログラムを開発中です。cuda - blockDim.xにアクセスできない？

問題は、blockDim.xとblockId.xにアクセスしようとすると、カーネルは常に起動して出力に失敗し、unknown errorが出力されます。

さらに、3x5の画像を使用した場合はthreadId.xにアクセスできますが、2048x2048の画像は使用できません。

PyCudaを使用するとカーネルコードが正常に実行されますが、現在はcuda Cに切り替える必要があります。

は、私は同じ設定がうまく機能し、問題が

私は配列ポインタを渡すとcudaMalloc
私のブロックサイズとグリッドサイズの設定（しかしに何か問題があるの方法に関連するかもしれないと思いますPyCudaだから私はそれを修正する方法を知らない）。

そして私はcuda-memcheckを使用し、私はunknown error 30を持って、私は解決策が、ノー役立つ情報については、Googleで検索。

私はカーネルで処理することができないので、何が問題になるのでしょうか？

EDIT

私は画像アレイは、1次元配列を「平坦化」とインデックス作成を行うと想定される意味1Dのインデックスを、使用したいです。

EDIT

私は、スレッドのチェックを追加した後、何か問題がまだあります。

__global__ void extractor(const unsigned char* in, unsigned char* out, int* debug) 
{ 
    int idx = (threadIdx.x) + blockDim.x * blockIdx.x ; 
    int y; int x; 
    int temp_x; int temp_y; int temp_idx; 
    int check = width*height; 
    if (idx < check) { 
     debug[0] = 1; // get kernel launch failed "unknown error" 
    } 
}

私はそれらの両方が同じエラーを取得し、スレッドのチェックブロックで、ブロック外の両方debug[0]=1;表現を置くことを試みました。

私はmemallocが正しく行われていないのではないかと疑いますか？

ところで、私はnvprofを使用し、それが

を言っ

=22344== Warning: Found 2 invalid records in the result. 
==22344== Warning: This can happen if device ran out of memory or if a device kernel was stopped due to an assertion.

EDIT

完全なコード：

#include "cuda_runtime.h" 
#include "device_launch_parameters.h" 

#include <cmath> 
#include <iostream> 

#include "PNG.h" 

#define L 3 
#define INC1 1 
#define INC2 1 
#define R_IN 2 
#define N_P 4 
#define BLOCK_SIZE 1024 
#define PI 3.14159265358979323846 

using namespace std; 

__global__ void extractor(const unsigned char* in, unsigned char* out, int* debug, int* disX, int* disY, int width, int height, int pad, int num_sample) 
{ 
    int idx = (threadIdx.x) + blockDim.x * blockIdx.x ; 
    int y; int x; 
    int temp_x; int temp_y; int temp_idx; 
    int check = width*height; 

     if (idx < check) { 
     debug[idx] = threadIdx.x; 
     y = idx/width; 
     x = idx%width; 
      if ((x < pad) || (x >= (width-pad)) || (y < pad) || (y >= (height-pad))) { 
       // need padding 
       for (int i = 0; i < num_sample; ++i){ 
        temp_x = x + disX[i]; 
        temp_y = y + disY[i]; 

        if (!((temp_x < 0)||(temp_x > (width-1)) || (temp_y < 0) ||(temp_y>(height-1)))) { 
        temp_idx = temp_y*width + temp_x; // sampled index 
        out[(idx*num_sample)+i] = in[temp_idx]; // copy sampled value to result 
       } 
       } 
      } else { 
       for (int i = 0; i < num_sample; ++i) 
       { 
        temp_x = x + disX[i]; 
        temp_y = y + disY[i]; 
        temp_idx = temp_y*width + temp_x; // sampled index 
        out[(idx*num_sample)+i] = in[temp_idx]; // copy sampled value to result 
       } 
      } 
     } 
    } 

vector<int> getCirclePos() { 
    int r = 0; 
    vector <int> circlePos; 
    while (!(r>(L/2))) { 
     circlePos.push_back(r); 
     if (r < R_IN) r += INC1; 
     else r += INC2; 
    } 
    cout << "circlePos:" << endl; 
    for (auto i = circlePos.begin(); i != circlePos.end(); ++i) 
    {cout << *i << ' ';} 
    cout << endl; 
    return circlePos; 
} 

int main(int arg, char* args[]) 
{ 
    cudaError_t cudaStatus; 
    vector<int> circlePos = getCirclePos(); 

    // get disX, disY 
    int num_sample_per_point = circlePos.size() * N_P; 
    int* disX = new int[num_sample_per_point]; 
    int* disY = new int[num_sample_per_point]; 
    int r; int cnt = 0; 
    for (int i = 0; i < circlePos.size(); ++i) 
    { 
     r = circlePos[i]; 
     float angle; 
     for (int j = 0; j < N_P; ++j) 
     { 
      angle = j*360.0/N_P; 
      disX[cnt] = r*cos(angle*M_PI/180.0); 
      disY[cnt] = r*sin(angle*M_PI/180.0); 
      // cout nvpro << disX[cnt] << "|" << disY[cnt]<< endl; 

      cnt++; 
     } 
    } 

    PNG inPng("test.png"); 
    // PNG outPng; 
    // outPng.Create(inPng.w, inPng.h); 

    //store width and height so we can use them for our output image later 
    const unsigned int w = inPng.w; 
    const unsigned int h = inPng.h; 
    cout << "w: " << w << " h: " << h << endl; 
    //4 because there are 4 color channels R, G, B, and A 
    int size = w * h; 

    unsigned char *in = 0; 
    unsigned char *out = 0; 
    int* debug = 0; 

    // Allocate GPU buffers for the images 
    cudaMalloc((void**)&in, size * sizeof(unsigned char)); 
    cudaMalloc((void**)&out, num_sample_per_point * size * sizeof(unsigned char)); 
    cudaMalloc((void**)&debug, size * sizeof(int)); 


    vector<unsigned char> img_data; 
    for (int i = 0; i < size; ++i) 
    { 
     img_data.push_back(inPng.data[i*4]); 
    } 

    // debug 
    cout << "========= img_data ==========" << endl; 
    for (int i = 0; i < size; ++i) 
    { 
     cout << int(img_data[i]) << "," ; 
    } 
    cout << endl; 

    // Copy image data from host memory to GPU buffers. 
    cudaMemcpy(in, &img_data[0], size * sizeof(unsigned char), cudaMemcpyHostToDevice); 

    //free the input image because we do not need it anymore 
    inPng.Free(); 

    // Launch a kernel on the GPU with one thread for each element. 
    dim3 b_dim(BLOCK_SIZE, 1, 1); // (1024, 1, 1) 
    dim3 g_dim(int(w*h/BLOCK_SIZE)+1, 1, 1); // (4097, 1, 1) 
    int pad = L/2; 

    // __global__ void extractor(const unsigned char* in, unsigned char* out, vector<int> disX, vector<int> disY, int width, int height, int pad, int num_sample) 
    extractor<<<g_dim, b_dim>>>(in, out, debug, disX, disY, w, h, pad, num_sample_per_point); 

    cudaStatus = cudaGetLastError(); 
    if (cudaStatus != cudaSuccess) 
    { 
     std::cout << "Kernel launch failed: " << cudaGetErrorString(cudaStatus) << std::endl; 
     cudaFree(in); 
     cudaFree(out); 
     cudaFree(debug); 
     exit(1); 
    } 

    auto tmp = new unsigned char[size*num_sample_per_point]; 
    auto tmp_debug = new int [size]; 

    cudaMemcpy(tmp_debug, debug, size * sizeof(int), cudaMemcpyDeviceToHost); 
    cudaMemcpy(tmp, out, num_sample_per_point * size * sizeof(unsigned char), cudaMemcpyDeviceToHost); 

    cout << "========= out =========" << endl; 
    for (int i = 0; i < size*num_sample_per_point; ++i) 
    { 
     cout << int(tmp[i]) << ", "; 
    } 
    cout << endl; 

    cout << "========debug=======" << endl; 
    for (int i = 0; i < size; ++i) 
    { 
     cout << tmp_debug[i] << ", "; 
    } 
    cout << endl; 

    cudaFree(in); 
    cudaFree(out); 
    cudaFree(debug); 

    delete[] tmp; delete[] tmp_debug; 

    return 0; 
}

出典

2017-06-03 Jinglei.Y

このブロックごとに1024個のスレッドを定義している（あなたのコメントによります）：

dim3 b_dim(BLOCK_SIZE, 1, 1); // (1024, 1, 1)

は、あなたの質問テキストによると、wとhが失敗した場合には、それぞれ2048年なので、この：

dim3 g_dim(int(w*h/BLOCK_SIZE)+1, 1, 1); // (4097, 1, 1)

は、あなたがあなたのコメントに示しているのと同様に、4097個のブロックを作成しています。

4097個のスレッドの合計は1024スレッドですが、割り当てサイズは合計で2048 * 2048個の要素または合計4194304個の要素しか提供していません。 4194324個のスレッドだけを4195328個のスレッドで起動し、1024個のスレッドを残しておきます。

これらの1024個の余分なスレッドは何をしますか？彼らはまだカーネルコードを実行し、割り当てられた領域の終わりを超えて配列debugにアクセスしようとします。

この結果、CおよびC++では動作が未定義になります。

この問題を解決するための慣習法はこのように、あなたのカーネルに問題のサイズを渡し、カーネルコード内で「スレッドのチェック」を追加することです：

やってから「エクストラ」のスレッドを防ぐ

__global__ void extractor(const unsigned char* in, unsigned char* out, int* debug, int n) 
{ 
    int idx = (threadIdx.x) + blockDim.x * blockIdx.x ; 
    if (idx < n) 
     debug[idx] = threadIdx.x; // debug variable is used for debugging 
}

何でも

ここでcudaタグで「スレッドチェック」を検索すると、このような質問の他の多くの例が見つかります。コード片に基づく例として

は、あなたが、私のためにエラーなしで、次のランを示している：あなたの完全なコードで

$ cat t147.cu 
const int width = 2048; 
const int height = 2048; 
const int BLOCK_SIZE = 1024; 
__global__ void extractor(const unsigned char* in, unsigned char* out, int* debug) 
{ 
    int idx = (threadIdx.x) + blockDim.x * blockIdx.x ; 
// int y; int x; 
// int temp_x; int temp_y; int temp_idx; 
    int check = width*height; 
    if (idx < check) { 
     debug[idx] = 1; // get kernel launch failed "unknown error" 
    } 
} 
int main(int arg, char* args[]) 
{ 

    const int w = width; 
    const int h = height; 
    const int num_sample_per_point = 1; 
    int size = w*h; // w is image width and h is image height 
    unsigned char *in = 0; 
    unsigned char *out = 0; 
    int* debug = 0; 

    // Allocate GPU buffers for the images 
    cudaMalloc((void**)&in, size * sizeof(unsigned char)); 
    cudaMalloc((void**)&out, num_sample_per_point * size * sizeof(unsigned char)); 
    cudaMalloc((void**)&debug, size * sizeof(int)); 

    // Copy image data from host memory to GPU buffers. 
// cudaMemcpy(in, &img_data[0], size * sizeof(unsigned char),cudaMemcpyHostToDevice); 

    dim3 b_dim(BLOCK_SIZE, 1, 1); // (1024, 1, 1) 
    dim3 g_dim(int(w*h/BLOCK_SIZE)+1, 1, 1); // (4097, 1, 1) 
    extractor<<<g_dim, b_dim>>>(in, out, debug); 
    cudaDeviceSynchronize(); 
} 
$ nvcc -arch=sm_61 -o t147 t147.cu 
$ cuda-memcheck ./t147 
========= CUDA-MEMCHECK 
========= ERROR SUMMARY: 0 errors 
$

、単にあなたのカーネルに不正アクセスの問題を抱えています。私はPNGへの依存を取り除くように修正しました。デバッグ設定以外のカーネルコードを省略するとうまく動作します。しかし、カーネルコードをインクルードしてcuda-memcheckで実行すると、あらゆる種類の境界外アクセスが発生します。将来的には、これらをデバッグするhereに記載の方法を使用することができる：

$ cat t146.cu 
#include <cmath> 
#include <iostream> 
#include <vector> 

#define L 3 
#define INC1 1 
#define INC2 1 
#define R_IN 2 
#define N_P 4 
#define BLOCK_SIZE 1024 
#define PI 3.14159265358979323846 

using namespace std; 

__global__ void extractor(const unsigned char* in, unsigned char* out, int* debug, int* disX, int* disY, int width, int height, int pad, int num_sample) 
{ 
    int idx = (threadIdx.x) + blockDim.x * blockIdx.x ; 
    int y; int x; 
    int temp_x; int temp_y; int temp_idx; 
    int check = width*height; 

     if (idx < check) { 
     debug[idx] = threadIdx.x; 
     y = idx/width; 
     x = idx%width; 
#ifdef FAIL 
      if ((x < pad) || (x >= (width-pad)) || (y < pad) || (y >= (height-pad))) { 
       // need padding 
       for (int i = 0; i < num_sample; ++i){ 
        temp_x = x + disX[i]; 
        temp_y = y + disY[i]; 

        if (!((temp_x < 0)||(temp_x > (width-1)) || (temp_y < 0) ||(temp_y>(height-1)))) { 
        temp_idx = temp_y*width + temp_x; // sampled index 
        out[(idx*num_sample)+i] = in[temp_idx]; // copy sampled value to result 
       } 
       } 
      } else { 
       for (int i = 0; i < num_sample; ++i) 
       { 
        temp_x = x + disX[i]; 
        temp_y = y + disY[i]; 
        temp_idx = temp_y*width + temp_x; // sampled index 
        out[(idx*num_sample)+i] = in[temp_idx]; // copy sampled value to result 
       } 
      } 
#endif 
     } 
    } 

vector<int> getCirclePos() { 
    int r = 0; 
    vector <int> circlePos; 
    while (!(r>(L/2))) { 
     circlePos.push_back(r); 
     if (r < R_IN) r += INC1; 
     else r += INC2; 
    } 
    cout << "circlePos:" << endl; 
    for (auto i = circlePos.begin(); i != circlePos.end(); ++i) 
    {//cout << *i << ' '; 
     } 
    cout << endl; 
    return circlePos; 
} 

int main(int arg, char* args[]) 
{ 
    cudaError_t cudaStatus; 
    vector<int> circlePos = getCirclePos(); 

    // get disX, disY 
    int num_sample_per_point = circlePos.size() * N_P; 
    int* disX = new int[num_sample_per_point]; 
    int* disY = new int[num_sample_per_point]; 
    int r; int cnt = 0; 
    for (int i = 0; i < circlePos.size(); ++i) 
    { 
     r = circlePos[i]; 
     float angle; 
     for (int j = 0; j < N_P; ++j) 
     { 
      angle = j*360.0/N_P; 
      disX[cnt] = r*cos(angle*M_PI/180.0); 
      disY[cnt] = r*sin(angle*M_PI/180.0); 
      // cout nvpro << disX[cnt] << "|" << disY[cnt]<< endl; 

      cnt++; 
     } 
    } 

    const unsigned int w = 2048; 
    const unsigned int h = 2048; 
    cout << "w: " << w << " h: " << h << endl; 
    //4 because there are 4 color channels R, G, B, and A 
    int size = w * h; 

    unsigned char *in = 0; 
    unsigned char *out = 0; 
    int* debug = 0; 

    // Allocate GPU buffers for the images 
    cudaMalloc((void**)&in, size * sizeof(unsigned char)); 
    cudaMalloc((void**)&out, num_sample_per_point * size * sizeof(unsigned char)); 
    cudaMalloc((void**)&debug, size * sizeof(int)); 


    vector<unsigned char> img_data; 
    for (int i = 0; i < size; ++i) 
    { 
     img_data.push_back(0); 
    } 

    // debug 
    cout << "========= img_data ==========" << endl; 
    for (int i = 0; i < size; ++i) 
    { 
//  cout << int(img_data[i]) << "," ; 
    } 
    cout << endl; 

    // Copy image data from host memory to GPU buffers. 
    cudaMemcpy(in, &img_data[0], size * sizeof(unsigned char), cudaMemcpyHostToDevice); 


    // Launch a kernel on the GPU with one thread for each element. 
    dim3 b_dim(BLOCK_SIZE, 1, 1); // (1024, 1, 1) 
    dim3 g_dim(int(w*h/BLOCK_SIZE)+1, 1, 1); // (4097, 1, 1) 
    int pad = L/2; 

    // __global__ void extractor(const unsigned char* in, unsigned char* out, vector<int> disX, vector<int> disY, int width, int height, int pad, int num_sample) 
    extractor<<<g_dim, b_dim>>>(in, out, debug, disX, disY, w, h, pad, num_sample_per_point); 

    cudaStatus = cudaGetLastError(); 
    if (cudaStatus != cudaSuccess) 
    { 
     std::cout << "Kernel launch failed: " << cudaGetErrorString(cudaStatus) << std::endl; 
     cudaFree(in); 
     cudaFree(out); 
     cudaFree(debug); 
     exit(1); 
    } 

    auto tmp = new unsigned char[size*num_sample_per_point]; 
    auto tmp_debug = new int [size]; 

    cudaMemcpy(tmp_debug, debug, size * sizeof(int), cudaMemcpyDeviceToHost); 
    cudaMemcpy(tmp, out, num_sample_per_point * size * sizeof(unsigned char), cudaMemcpyDeviceToHost); 

    cout << "========= out =========" << endl; 
    for (int i = 0; i < size*num_sample_per_point; ++i) 
    { 
    //  cout << int(tmp[i]) << ", "; 
    } 
    cout << endl; 

    cout << "========debug=======" << endl; 
    for (int i = 0; i < size; ++i) 
    { 
    // cout << tmp_debug[i] << ", "; 
    } 
    cout << endl; 

    cudaFree(in); 
    cudaFree(out); 
    cudaFree(debug); 

    delete[] tmp; delete[] tmp_debug; 

    return 0; 
} 
$ nvcc -std=c++11 -o t146 t146.cu -arch=sm_61 -lineinfo 
t146.cu(18): warning: variable "y" was set but never used 

t146.cu(18): warning: variable "x" was set but never used 

t146.cu(19): warning: variable "temp_x" was declared but never referenced 

t146.cu(19): warning: variable "temp_y" was declared but never referenced 

t146.cu(19): warning: variable "temp_idx" was declared but never referenced 

t146.cu(18): warning: variable "y" was set but never used 

t146.cu(18): warning: variable "x" was set but never used 

t146.cu(19): warning: variable "temp_x" was declared but never referenced 

t146.cu(19): warning: variable "temp_y" was declared but never referenced 

t146.cu(19): warning: variable "temp_idx" was declared but never referenced 

$ cuda-memcheck ./t146 
========= CUDA-MEMCHECK 
circlePos: 

w: 2048 h: 2048 
========= img_data ========== 

========= out ========= 

========debug======= 

========= ERROR SUMMARY: 0 errors 
$ nvcc -std=c++11 -o t146 t146.cu -arch=sm_61 -lineinfo -DFAIL 
$ cuda-memcheck ./t146 
... 
========= Invalid __global__ read of size 4 
=========  at 0x00000418 in /home/ubuntu/bobc/misc/t146.cu:41:extractor(unsigned char const *, unsigned char*, int*, int*, int*, int, int, int, int) 
=========  by thread (197,0,0) in block (17,0,0) 
=========  Address 0x00c8b290 is out of bounds 
=========  Saved host backtrace up to driver entry point at kernel launch time 
=========  Host Frame:/usr/lib/x86_64-linux-gnu/libcuda.so.1 (cuLaunchKernel + 0x2c5) 
... 
(and much more output like this)

上記出力点がdisXから読み出しているコード、41行目に。

結局のところ、あなたのdisXがホストに割り当てられた変数です：

int* disX = new int[num_sample_per_point];

いますが、デバイスコードに渡ししようとしている。

extractor<<<g_dim, b_dim>>>(in, out, debug, disX, disY, w, h, pad, num_sample_per_point); 
              ^^^^

だけで完全に壊れています。あなたはCUDAでそれをすることはできません。あなたは、その変数のデバイスのコピーを作成する必要がある、と私はその問題を解決する場合もdisY、変更されたコードは私のために、エラーなしで実行：返信用

$ cat t146.cu 
#include <cmath> 
#include <iostream> 
#include <vector> 

#define L 3 
#define INC1 1 
#define INC2 1 
#define R_IN 2 
#define N_P 4 
#define BLOCK_SIZE 1024 
#define PI 3.14159265358979323846 

using namespace std; 

__global__ void extractor(const unsigned char* in, unsigned char* out, int* debug, int* disX, int* disY, int width, int height, int pad, int num_sample) 
{ 
    int idx = (threadIdx.x) + blockDim.x * blockIdx.x ; 
    int y; int x; 
    int temp_x; int temp_y; int temp_idx; 
    int check = width*height; 

     if (idx < check) { 
     debug[idx] = threadIdx.x; 
     y = idx/width; 
     x = idx%width; 
#ifdef FAIL 
      if ((x < pad) || (x >= (width-pad)) || (y < pad) || (y >= (height-pad))) { 
       // need padding 
       for (int i = 0; i < num_sample; ++i){ 
        temp_x = x + disX[i]; 
        temp_y = y + disY[i]; 

        if (!((temp_x < 0)||(temp_x > (width-1)) || (temp_y < 0) ||(temp_y>(height-1)))) { 
        temp_idx = temp_y*width + temp_x; // sampled index 
        out[(idx*num_sample)+i] = in[temp_idx]; // copy sampled value to result 
       } 
       } 
      } else { 
       for (int i = 0; i < num_sample; ++i) 
       { 
        temp_x = x + disX[i]; 
        temp_y = y + disY[i]; 
        temp_idx = temp_y*width + temp_x; // sampled index 
        out[(idx*num_sample)+i] = in[temp_idx]; // copy sampled value to result 
       } 
      } 
#endif 
     } 
    } 

vector<int> getCirclePos() { 
    int r = 0; 
    vector <int> circlePos; 
    while (!(r>(L/2))) { 
     circlePos.push_back(r); 
     if (r < R_IN) r += INC1; 
     else r += INC2; 
    } 
    cout << "circlePos:" << endl; 
    for (auto i = circlePos.begin(); i != circlePos.end(); ++i) 
    {//cout << *i << ' '; 
     } 
    cout << endl; 
    return circlePos; 
} 

int main(int arg, char* args[]) 
{ 
    cudaError_t cudaStatus; 
    vector<int> circlePos = getCirclePos(); 

    // get disX, disY 
    int num_sample_per_point = circlePos.size() * N_P; 
    int* disX = new int[num_sample_per_point]; 
    int* disY = new int[num_sample_per_point]; 
    int r; int cnt = 0; 
    for (int i = 0; i < circlePos.size(); ++i) 
    { 
     r = circlePos[i]; 
     float angle; 
     for (int j = 0; j < N_P; ++j) 
     { 
      angle = j*360.0/N_P; 
      disX[cnt] = r*cos(angle*M_PI/180.0); 
      disY[cnt] = r*sin(angle*M_PI/180.0); 
      // cout nvpro << disX[cnt] << "|" << disY[cnt]<< endl; 

      cnt++; 
     } 
    } 

    int *d_disX, *d_disY; 
    cudaMalloc(&d_disX, num_sample_per_point*sizeof(int)); 
    cudaMalloc(&d_disY, num_sample_per_point*sizeof(int)); 
    cudaMemcpy(d_disX, disX, num_sample_per_point*sizeof(int), cudaMemcpyHostToDevice); 
    cudaMemcpy(d_disY, disY, num_sample_per_point*sizeof(int), cudaMemcpyHostToDevice); 
    const unsigned int w = 2048; 
    const unsigned int h = 2048; 
    cout << "w: " << w << " h: " << h << endl; 
    //4 because there are 4 color channels R, G, B, and A 
    int size = w * h; 

    unsigned char *in = 0; 
    unsigned char *out = 0; 
    int* debug = 0; 

    // Allocate GPU buffers for the images 
    cudaMalloc((void**)&in, size * sizeof(unsigned char)); 
    cudaMalloc((void**)&out, num_sample_per_point * size * sizeof(unsigned char)); 
    cudaMalloc((void**)&debug, size * sizeof(int)); 


    vector<unsigned char> img_data; 
    for (int i = 0; i < size; ++i) 
    { 
     img_data.push_back(0); 
    } 

    // debug 
    cout << "========= img_data ==========" << endl; 
    for (int i = 0; i < size; ++i) 
    { 
//  cout << int(img_data[i]) << "," ; 
    } 
    cout << endl; 

    // Copy image data from host memory to GPU buffers. 
    cudaMemcpy(in, &img_data[0], size * sizeof(unsigned char), cudaMemcpyHostToDevice); 


    // Launch a kernel on the GPU with one thread for each element. 
    dim3 b_dim(BLOCK_SIZE, 1, 1); // (1024, 1, 1) 
    dim3 g_dim(int(w*h/BLOCK_SIZE)+1, 1, 1); // (4097, 1, 1) 
    int pad = L/2; 

    // __global__ void extractor(const unsigned char* in, unsigned char* out, vector<int> disX, vector<int> disY, int width, int height, int pad, int num_sample) 
    extractor<<<g_dim, b_dim>>>(in, out, debug, d_disX, d_disY, w, h, pad, num_sample_per_point); 

    cudaStatus = cudaGetLastError(); 
    if (cudaStatus != cudaSuccess) 
    { 
     std::cout << "Kernel launch failed: " << cudaGetErrorString(cudaStatus) << std::endl; 
     cudaFree(in); 
     cudaFree(out); 
     cudaFree(debug); 
     exit(1); 
    } 

    auto tmp = new unsigned char[size*num_sample_per_point]; 
    auto tmp_debug = new int [size]; 

    cudaMemcpy(tmp_debug, debug, size * sizeof(int), cudaMemcpyDeviceToHost); 
    cudaMemcpy(tmp, out, num_sample_per_point * size * sizeof(unsigned char), cudaMemcpyDeviceToHost); 

    cout << "========= out =========" << endl; 
    for (int i = 0; i < size*num_sample_per_point; ++i) 
    { 
    //  cout << int(tmp[i]) << ", "; 
    } 
    cout << endl; 

    cout << "========debug=======" << endl; 
    for (int i = 0; i < size; ++i) 
    { 
    // cout << tmp_debug[i] << ", "; 
    } 
    cout << endl; 

    cudaFree(in); 
    cudaFree(out); 
    cudaFree(debug); 

    delete[] tmp; delete[] tmp_debug; 

    return 0; 
} 
$ nvcc -std=c++11 -o t146 t146.cu -arch=sm_61 -lineinfo -DFAIL 
$ cuda-memcheck ./t146 
========= CUDA-MEMCHECK 
circlePos: 

w: 2048 h: 2048 
========= img_data ========== 

========= out ========= 

========debug======= 

========= ERROR SUMMARY: 0 errors 
$

出典

2017-06-03 15:20:30

こんにちは、感謝を。あなたが言ったようにコードを変更しましたが、スレッドチェックを行ってもカーネルの 'debug'変数にアクセスすることはできません。私の編集を見てください。 –

あなたは[mcve]を提供する必要があります。 'width'と' height'が何であるか分かりません。それは何かを追加したり何も変更しなくてもコピー＆ペーストやコンパイルができる*完全なコードでなければなりません。 –

質問の詳細に追加されました。前もって感謝します。 –

cuda - blockDim.xにアクセスできない？

答えて

関連する問題