Проблемы с 2D массивом CUDA

У вас есть две ошибки

1. Каждый поток в ядре должен выполнять одну операцию, а не все операции. (По соображениям памяти вы можете захотеть сделать больше, но мы оставим этот пример простым).

2. Когда вы загружали данные на устройство, вы переключали высоту тона назначения и источника.

Вот рабочая версия

#include <cuda_runtime.h>
#include <stdlib.h>
#include <iostream>
#include <sstream> 

#define CUDASAFECALL( err ) cuda_safe_call(err, __FILE__, __LINE__ )
void cuda_safe_call(const cudaError err, const char *file, const int line)
{
    if (cudaSuccess != err)
    {
        std::stringstream error_msg;
        error_msg << "cuda_safe_call() failed at " << file << ":" << line << ":" << cudaGetErrorString(err);
        const auto error_msg_str = error_msg.str();
        std::cout << error_msg_str << std::endl;
        throw std::runtime_error(error_msg_str);
    }
}

__global__ void add_d2D(const double *x, const double *y, double *z, int n, int m, int m_pitch_elements) 
{
    int row = blockIdx.x * blockDim.x + threadIdx.x;
    int col = blockIdx.y * blockDim.y + threadIdx.y;
    if (row< n && col <m )
    {
        auto idx = row*m_pitch_elements + col;
        z[idx] = x[idx] + y[idx];
        //z[idx] = idx;
    }
}

__host__ void add2D(const double *a,const double *b, double *result, int N, int M) {
    double *a_d, *b_d, *c_d;
    size_t pitcha,pitchb,pitchc;

    CUDASAFECALL(cudaMallocPitch(&a_d, &pitcha, M * sizeof(double), N));
    CUDASAFECALL(cudaMallocPitch(&b_d, &pitchb, M * sizeof(double), N));
    CUDASAFECALL(cudaMallocPitch(&c_d, &pitchc, M * sizeof(double), N));
    CUDASAFECALL(cudaMemcpy2D(a_d, pitcha, a, M * sizeof(double), M * sizeof(double), N, cudaMemcpyHostToDevice));
    CUDASAFECALL(cudaMemcpy2D(b_d, pitchb, b, M * sizeof(double), M * sizeof(double), N, cudaMemcpyHostToDevice));

    dim3 threadsPerBlock(2, 2);
    auto safediv = [](auto a, auto b) {return static_cast<unsigned int>(ceil(a / (b*1.0))); };
    dim3 numBlocks(safediv(N, threadsPerBlock.x), safediv( M, threadsPerBlock.y));
    //all the pitches should be the same
    auto pitch_elements = pitcha / sizeof(double);
    add_d2D << <numBlocks, threadsPerBlock >> >(a_d, b_d, c_d, N, M, pitch_elements);

    CUDASAFECALL(cudaDeviceSynchronize());

    CUDASAFECALL(cudaMemcpy2D(result, M * sizeof(double), c_d, pitchc, M * sizeof(double), N, cudaMemcpyDeviceToHost));

    CUDASAFECALL(cudaFree(a_d));
    CUDASAFECALL(cudaFree(b_d));
    CUDASAFECALL(cudaFree(c_d));
}

int main()
{
    double a[4][10];
    double b[4][10];
    double c[4][10];
    for (int i = 0; i < 4; i++) {
        for (int j = 0; j < 10; j++) {
            a[i][j] = 0 + rand() % 10;
            b[i][j] = 0 + rand() % 10;
        }
    }
    add2D((double *)a, (double *)b, (double *)c, 4, 10);
    for (int i = 0; i < 4; i++) {
        for (int j = 0; j < 10; j++) {
            std::cout << a[i][j] << " " << b[i][j] << " " << c[i][j]<< "|"<< a[i][j]+ b[i][j] << std::endl;
        }
    }
    return 0;
}