Понимание индексации CUDA - PullRequest
Новая
       19

Понимание индексации CUDA

0 голосов
/ 04 апреля 2020

Я унаследовал некоторый код CUDA, над которым мне нужно поработать, но некоторая индексация, выполненная в нем, сбивает меня с толку.

Простым примером была бы нормализация данных. Скажем, у нас есть общий массив A[2*N], который представляет собой матрицу формы 2xN, развернутую в массив. Тогда у нас есть средства нормализации и стандартное отклонение: norm_means[2] и norm_stds[2]. Цель состоит в том, чтобы нормализовать данные в A параллельно. Минимальный пример: 

__global__ void normalise(float *data, float *norm, float *std) {
int tdy = threadIdx.y;

for (int i=tdy; i<D; i+=blockDim.y)
  data[i] = data[i] * norm[i] + std[i];
}

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

// generate data
int N=100;
int D=2;
MatrixXd A = MatrixXd::Random(N*D,1);
MatrixXd norm_means = MatrixXd::Random(D,1);
MatrixXd norm_stds = MatrixXd::Random(D,1);

// transfer data to device
float* A_d;
float* norm_means_d;
float* nrom_stds_d;
cudaMalloc((void **)&A_d, N * D * sizeof(float));
cudaMalloc((void **)&norm_means_d, D * sizeof(float));
cudaMalloc((void **)&norm_stds_d, D * sizeof(float));
cudaMemcpy(A_d, A.data(), D * N * sizeof(float), cudaMemcpyHostToDevice);
cudaMemcpy(norm_means_d, norm_means.data(), D * sizeof(float), cudaMemcpyHostToDevice);
cudaMemcpy(norm_stds_d, norm_stds.data(), D * sizeof(float), cudaMemcpyHostToDevice);

// Setup execution
const int BLOCKSIZE_X = 8;
const int BLOCKSIZE_Y = 16;
const int GRIDSIZE_X = (N-1)/BLOCKSIZE_X + 1;
dim3 dimBlock(BLOCKSIZE_X, BLOCKSIZE_Y, 1);
dim3 dimGrid(GRIDSIZE_X, 1, 1);

normalise<<dimGrid, dimBlock, 0>>>(A_d, norm_means_d, norm_stds_d);

}

Обратите внимание, что я использую Eigen для генерации матрицы. Для краткости я пропустил включения:

Этот код, приведенный выше, через некоторые маги c работает и достигает желаемых результатов. Однако функция ядра CUDA не имеет для меня никакого смысла, потому что for l oop должен останавливаться после одного выполнения как i>D после первой итерации ... но это не так?

Если я изменю ядро, которое имеет для меня больше смысла, например. 

__global__ void normalise(float *data, float *norm, float *std) {
int tdy = threadIdx.y;
for (int i=0; i<D; i++)
  data[tdy + i*blockDim.y] = data[tdy + i*blockDim.y] * norm[i] + std[i];
}

, программа перестает работать и просто выводит данные gibberi sh.

Может кто-нибудь объяснить, почему у меня такое поведение?

PS. Я очень новичок в CUDA

1 Ответ

2 голосов
/ 04 апреля 2020

Действительно, бессмысленно иметь двумерное ядро ​​для выполнения поэлементной операции над массивом. Также нет причин работать в блоках размером 8х16. Но ваше модифицированное ядро ​​использует измерение second (y) only ; Вероятно, поэтому это не работает. Возможно, вам нужно было использовать только измерение first (x).

Однако - было бы разумно разумно использовать измерение Y для фактического второго измерения, например что-то как это: 

__global__ void normalize(
          float __restrict *data,
    const float __restrict *norm,
    const float __restrict *std) 
{
    auto pos = threadIdx.x + blockDim.x * blockIdx.x;
    auto d = threadIdx.y + blockDim.y * blockIdx.y; // or maybe just threadIdx.y;
    data[pos + d * N] = data[pos + d * N] * norm[d] + std[d];
}

Другие вопросы для рассмотрения:

  • Я добавил __restrict к вашим указателям. Всегда делать это при необходимости; вот почему .
  • Хорошая идея - иметь один поток для работы с более чем одним элементом данных - но вы должны сделать это в более длинном измерении, где поток может использовать повторно. его нормальные и стандартные значения, а не читать их из памяти каждый раз.
...