Передача динамического двумерного массива в kernal и на host. cuda

Question

Передача динамического двумерного массива в kernal и на host. cuda

Рейтинг: -1Ответов: 1Опубликовано: 23.03.2023

есть код, который должен передавать матрицы в ядро, складывать их и возвращать обратно на хост матрицу с ответом(ответ сложения матриц записывается в третью матрицу). Передать на устройство и обратно у меня с горем пополам получилось, а вот в самом ядре, выполнение программы просто не доходит до условия(пробовал без условия, всё равно не присваивает значения. Также проверял принтами, программа просто не доходит до строки с присваиванием). В чём проблема и как это исправить? Главное условие, чтобы матрица была создана с помощью двумерного динамического массива с использованием malloc(без new), а доступ к элементам матрицы в ядре осуществлялся с помощью оператора двойных скобок. В общем за это задание обещали автомат =) Сам уже второй день голову ломаю. Буду очень благодарен, если кто-то подскажет, хотя бы в каком направлении мыслить. Код:

#include <iostream>

using namespace std;
#define n 800

__global__ void matAdd(float**, float**,
                       float**, int);

int main()
{
    // Выделение памяти для хост матриц
    float** a = (float**)malloc(n * sizeof(float*));
    float** b = (float**)malloc(n * sizeof(float*));
    float** c = (float**)malloc(n * sizeof(float*));

    for (int i = 0; i < n; i++) {
        a[i] = (float*)malloc(n * sizeof(float));
        b[i] = (float*)malloc(n * sizeof(float));
        c[i] = (float*)malloc(n * sizeof(float));

        // Инициализация двух хост матриц
        for (int j = 0; j < n; j++) {
            a[i][j] = b[i][j] = 0.5;
            c[i][j] = 0;
        }
    }

    // Создание девайс матриц
    float** devA;
    float** devB;
    float** devC;
    size_t pitch;

    // Выделение памяти на устройстве
    cudaMallocPitch((void**)&devA, &pitch, n * sizeof(float), n);
    cudaMallocPitch((void**)&devB, &pitch, n * sizeof(float), n);
    cudaMallocPitch((void**)&devC, &pitch, n * sizeof(float), n);

    // Копирование данных на устройство
    cudaMemcpy2D(devA, pitch, a, n * sizeof(float), n * sizeof(float), n, 
cudaMemcpyHostToDevice);
    cudaMemcpy2D(devB, pitch, b, n * sizeof(float), n * sizeof(float), n, 
cudaMemcpyHostToDevice);

    // Размерность блоков и количество нитей
    dim3 numThreadsPerBlock(10, 10);
    dim3 numBlocks((n + numThreadsPerBlock.x - 1) / numThreadsPerBlock.x,
                    (n + numThreadsPerBlock.y - 1) / numThreadsPerBlock.y);
    matAdd<<<numBlocks, numThreadsPerBlock>>>(devA, devB, devC,
                                              numThreadsPerBlock.x);

    // Копирование финальной матрицы на хост
    cudaMemcpy2D(c, n * sizeof(float), devC, pitch, n * sizeof(float), n, 
cudaMemcpyDeviceToHost);

    cudaFree(devA);
    cudaFree(devB);
    cudaFree(devC);

    for (int i = 0; i < 5; i++) {
        for (int j = 0; j < 5; j++) {
            cout << c[i][j] << ' ';
        }
        cout << '\n';
    }

    for (int i = 0; i < n; i++) {
        free(a[i]);
        free(b[i]);
        free(c[i]);
    }
    free(a);
    free(b);
    free(c);

    return 0;   
}

__global__ void matAdd(float** A, float** B,
                       float** C, int step)
{
    int i = step * blockIdx.x + threadIdx.x;
    int j = step * blockIdx.y + threadIdx.y;

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

c# cuda

Источник: Stack Overflow на русском

Answer 1

▲ 0

В общем нашёл я таки решение. Не могу точно сказать, почему, но куда плохо работает с двойным указателем через операцию две звёздочки (**). Рабочий прилагаю нише

    #include <iostream>

    using namespace std;
    #define n 800

    __global__ void matAdd(float (*)[n], float (*)[n],
                           float (*)[n]);

int main()
{
    const int memSize = sizeof(float) * n * n;
    // Выделение памяти для хост матриц
    float (*a)[n] = (float(*)[n])malloc(memSize);
    float (*b)[n] = (float(*)[n])malloc(memSize);
    float (*c)[n] = (float(*)[n])malloc(memSize);

        // Инициализация двух хост матриц
    for (int i = 0; i < n; i++) {
        for (int j = 0; j < n; j++) {
            a[i][j] = b[i][j] = 0.5;
            c[i][j] = 0;
        }
    }

    // Создание девайс матриц
    float (*devA)[n];
    float (*devB)[n];
    float (*devC)[n];
    size_t pitch;

    // Выделение памяти на устройстве
    cudaMallocPitch(&devA, &pitch, n * sizeof(float), n);
    cudaMallocPitch(&devB, &pitch, n * sizeof(float), n);
    cudaMallocPitch(&devC, &pitch, n * sizeof(float), n);

    cudaMemcpy2D(devA, pitch, a, n * sizeof(float), n * sizeof(float), n, cudaMemcpyHostToDevice);
    cudaMemcpy2D(devB, pitch, b, n * sizeof(float), n * sizeof(float), n, cudaMemcpyHostToDevice);

    dim3 numThreadsPerBlock(10, 10);
    dim3 numBlocks((n + numThreadsPerBlock.x - 1) / numThreadsPerBlock.x,
                    (n + numThreadsPerBlock.y - 1) / numThreadsPerBlock.y);
    matAdd<<<numBlocks, numThreadsPerBlock>>>(devA, devB, devC);

    cudaMemcpy2D(c, n * sizeof(float), devC, pitch, n * sizeof(float), n, cudaMemcpyDeviceToHost);

    cudaFree(devA);
    cudaFree(devB);
    cudaFree(devC);

    for (int i = 0; i < 5; i++) {
        for (int j = 0; j < 5; j++) {
            cout << c[i][j] << ' ';
        }
        cout << '\n';
    }

    free(a);
    free(b);
    free(c);

    return 0;   
}

__global__ void matAdd(float (*A)[n], float (*B)[n],
                       float (*C)[n])
{
    int i = blockDim.x * blockIdx.x + threadIdx.x;
    int j = blockDim.y * blockIdx.y + threadIdx.y;
    
    if (i < n && j < n) 
        C[i][j] = A[i][j] + B[i][j];
}

Передача динамического двумерного массива в kernal и на host. cuda

Ответы