CUDA：在兩個維度上訪問任意長矩陣

Question

嘿， 我目前只用在一個維度索引線程訪問這樣的一個矩陣的所有元素：

// Thread-ID 
int idx = blockIdx.x * blockDim.x + threadIdx.x; 
// Offset: 
int offset = gridDim.x * blockDim.x; 

while(idx < MATRIX_ROWS * MATRIX_COLS) 
{ 
    row = idx % MATRIX_ROWS; 
    col = idx/MATRIX_ROWS;  

    matrix[ row ][ col ] = ...; 
    idx += offset; 
} 

現在我不知道如何訪問具有二維索引的任意長矩陣。我希望一個塊總是訪問一行中的單個元素。類似的東西（X-指數指的cols和在y索引到矩陣的行）：

// Thread-IDs 
int idx = blockIdx.x * blockDim.x + threadIdx.x; 
int idy = blockIdx.y * blockDim.y + threadIdx.y; 

// Offset: 
int offset = gridDim.x * blockDim.x; 

while(idx < MATRIX_COLS) 
{ 
    matrix[ idy ][ idx ] = ...; 
    idx += offset; 
} 

現在讓我們假設矩陣具有比我調用內核時開始塊多行：當開始N塊，矩陣的前N行處理正確，但其他行呢？你會怎麼做？ 謝謝！

編輯：我想出了一個想法，但我不知道這是不知何故'難看'編碼！？

// Thread-IDs 
int idx0 = blockIdx.x * blockDim.x + threadIdx.x; 
int idx = idx0; 
int idy = blockIdx.y * blockDim.y + threadIdx.y; 

// Offset: 
int offsetx = gridDim.x * blockDim.x; 
int offsety = gridDim.y * blockDim.y; 

while(idx < MATRIX_COLS && idy < MATRIX_ROWS) 
{ 
    matrix[ idy ][ idx ] = ...; 

    idx += offsetx; 
    if(idx > MATRIX_COLS) 
    { 
     // Jump to nex row and start from 'beginning' concerning columns 
     idy += offsety; 
     idx = idx0; 
    } 
} 

Answer 1

也許像這樣的東西是你在找什麼？

// Thread-IDs 
int idx = blockIdx.x * blockDim.x + threadIdx.x; 
int idy = blockIdx.y * blockDim.y + threadIdx.y; 

// Offset: 
int offsetx = gridDim.x * blockDim.x; 
int offsety = gridDim.y * blockDim.y; 

for(row = idy; row < MATRIX_ROWS; i+=offsety) { 
    float * pos = matrix + row; 

#pragma unroll 
    for(col = idx; col < MATRIX_COLS; col+=offsetx) { 
     pos[col] = .....; 
    } 
} 

如果MATRIX_COLS被預處理器定義或常量，編譯器或許能解開內循環，並給予多一點的表現。

編輯：代碼的第一個版本是寫入列主要順序卡在我的頭後面，所以忽略這裏的評論。應該通過這種方式合併訪問。