opencv - cudaカーネルのopencv 3チャネル画像データオフセット

Question

次のコード スニペットがあります。

// main.cu

#define thread 16

dim3 blocks( ( width + thread - 1 ) / thread, ( height + thread - 1 ) / thread );

dim3 threads( thread, thread );

kernel<<<blocks, threads>>>( dev_data, width, height );

// kernel function

__global__ void kernel( uchar *data, int width, int height ) {

// map from threadIdx/blockIdx to pixel position

int x = threadIdx.x + blockIdx.x * blockDim.x;

int y = threadIdx.y + blockIdx.y * blockDim.y;

int offset = x + y * blockDim.x * gridDim.x;

data[offset * 3 + 0] = 255;

data[offset * 3 + 1] = 0;

data[offset * 3 + 2] = 0;
}

実行すると、ピクセルの半分近くだけが青くなります。私は何を間違っていますか？インデックス作成についてだと思います。