операции над двумерным массивом в ядре CUDA для matlab

Question

предположим, у меня есть следующий серийный номер C:

int add(int* a, int* b, int n)
{
    for(i=0; i<n; i++)
    {
        for(j=0; j<n; j++)
        {
            a[i][j]+=b[i][j];
        }
    }

    return 0;
}

Я думаю, что лучший способ паралелизировать это понять, что это проблема 2D, и использовать блоки 2D потоков согласно ядру CUDA - вложено в цикл

Имея это в виду, я начал писать свой cuda kernal следующим образом:

__global__ void calc(int **A, int **B, int n)
{

    int i= blockIdx.x * blockDim.x + threadIdx.x;
    int j= blockIdx.y * blockDim.y + threadIdx.y;


    if (i>=n || j>=n)
        return;

    A[i][j]+=B[i][j];


}

nvcc говорит мне, что:

./addm.cu(13): Warning: Cannot tell what pointer points to, assuming global memory space
./addm.cu(13): Warning: Cannot tell what pointer points to, assuming global memory space
./addm.cu(13): Warning: Cannot tell what pointer points to, assuming global memory space  

1) Я правильно со своей философией? 2) Я думаю, что понимаю блоки, потоки и т. Д., Но я не понимаю, что

    int i= blockIdx.x * blockDim.x + threadIdx.x;
    int j= blockIdx.y * blockDim.y + threadIdx.y;

делает

3) Это самый эффективный / самый быстрый способ выполнения операций над двумерным массивом в целом? т.е. не просто добавление матрицы, это может быть любая операция «элемент за элементом».

4) Смогу ли я позвонить с Matlab? обычно это уродство, когда прототип имеет форму type** var

Спасибо, ребята

Answer 1

Предупреждения компилятора, которые вы получаете, связаны с тем, что на старых графических процессорах структура памяти не является «плоской».Компилятор не может знать, в каком пространстве памяти находятся адреса, удерживаемые массивами указателей, в которых работает ваше ядро.Поэтому он предупреждает вас, что он предполагает, что операция выполняется в глобальной памяти.Если вы скомпилируете код для карты Fermi (архитектура sm_20 или sm_21), вы не увидите этого предупреждения, потому что модель памяти на этих картах «плоская», а указатели правильно интерпретируются аппаратным обеспечением во время выполнения.Компилятору не нужно обрабатывать его во время компиляции.

Чтобы ответить на каждый из ваших вопросов:

1. Да.И нет.Общая идея верна примерно на 90%, но есть несколько проблем реализации, которые станут очевидными из ответов, которые следуют.

2. CUDA C имеет встроенные переменные, позволяющие каждому потоку определять свои«координаты» в исполнительной сетке, в которой он запущен, а также размеры каждого блока и сетки.threadIdx.{xyz} предоставляет координаты потока внутри блока, а blockIdx.{xyz} координату блока с сеткой.blockDim.{xyz} и gridDim.{xyz} предоставляют размеры блока и сетки соответственно (обратите внимание, что не все аппаратные средства поддерживают 3D-сетки).CUDA использует основной порядок столбцов для нумерации потоков в каждом блоке и блока в каждой сетке.Запрашиваемый вами расчет вычисляет эквивалентную {i,j} координату в двумерной сетке, используя координаты потока и блока и размер блока.Это обсуждается более подробно на первых страницах главы «Модель программирования» руководства по программированию CUDA.

3. Нет, и я говорю это по двум причинам.

   Во-первых, использование массивов указателей для доступа к памяти не является хорошей идеей в CUDA.Два уровня косвенного обращения к указателю значительно увеличивают штраф за задержку при получении ваших данных.Ключевым отличием типичной архитектуры графического процессора от современной архитектуры процессора является система памяти.Графические процессоры имеют потрясающе высокую пиковую пропускную способность памяти, но очень высокую задержку доступа, тогда как центральные процессоры рассчитаны на минимальную задержку.Таким образом, необходимость чтения и косвенного двух указателей для извлечения значения из памяти является очень большим снижением производительности.Вместо этого сохраните 2D-массив или матрицу в линейной памяти.Это то, что BLAS, LAPACK и Matlab делают в любом случае.

   Во-вторых, каждый поток в вашем коде выполняет четыре целочисленных арифметических операции с затратами на установку (вычисления индекса) для каждой "целой" производной целочисленной операции (сложение),Существуют стратегии, позволяющие уменьшить это, обычно включающие в себя обработку каждого потока более чем одним элементом массива.

   Если бы я писал ядро ​​для этой операции, я сделал бы это как код в нижней части моего ответа.При этом используется линейная память и 1D сетка .Подходящее количество потоков, чтобы должным образом занимать графический процессор, обрабатывает весь входной массив, причем каждый поток обрабатывает множество входных данных.

4. Нет.Как я упоминал ранее в своем ответе, Matlab использует линейную память для хранения матриц, а не массив указателей.Это не соответствует макету, который ожидает ваш код ядра.

Пример кода:

__global__ void calc(int *A, int *B, int N)
{

    int i = blockIdx.x * blockDim.x + threadIdx.x;
    int s = blockDim.x * gridDim.x;

    for( ; i<N; i+=s) {
        A[i] += B[i];
    }
}

Answer 2

Я предполагаю, что вы работаете с n-на-n массивом основных порядков строк.Попробуйте следующее:

__global__ void calc(int *A, int *B, int n)
{
    int i= blockIdx.x * blockDim.x + threadIdx.x;
    int j= blockIdx.y * blockDim.y + threadIdx.y;

    if (i<n && j<n) {
        A[i*n+j] += B[i*n+j];
    }
}