Для справки, это домашнее задание, так что помогите как можно меньше или столько же, помня об этом. Мы используем постоянную память для хранения «маскирующей матрицы», которая будет использоваться для выполнения свертки на большей матрице. Когда я нахожусь в коде хоста, я копирую маску в постоянную память с помощью cudaMemcpyToSymbol ().

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

Должно ли объявление постоянной памяти быть включено в отдельный файл kernel.cu?

Я минимизирую код, чтобы просто показать, что связано с постоянной памятью. Как таковой, пожалуйста, не указывайте, если что-то не инициализировано и т. Д. Код для этого есть, но сейчас это не вызывает беспокойства.

main.cu:

#include <stdio.h>
#include "kernel.cu"

__constant__ float M_d[FILTER_SIZE * FILTER_SIZE];

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

     Matrix M_h, N_h, P_h; // M: filter, N: input image, P: output image

    /* Allocate host memory */
    M_h = allocateMatrix(FILTER_SIZE, FILTER_SIZE);
    N_h = allocateMatrix(imageHeight, imageWidth);
    P_h = allocateMatrix(imageHeight, imageWidth);

    /* Initialize filter and images */
    initMatrix(M_h);
    initMatrix(N_h);


    cudaError_t cudda_ret = cudaMemcpyToSymbol(M_d, M_h.elements, M_h.height * M_h.width * sizeof(float), 0, cudaMemcpyHostToDevice);
    //char* cudda_ret_pointer = cudaGetErrorString(cudda_ret);

    if( cudda_ret != cudaSuccess){
        printf("\n\ncudaMemcpyToSymbol failed\n\n");
        printf("%s, \n\n", cudaGetErrorString(cudda_ret));
    }


    // Launch kernel ----------------------------------------------------------
    printf("Launching kernel..."); fflush(stdout);

    //INSERT CODE HERE
    //block size is 16x16
    //              \\\\\\\\\\\\\**DONE**
    dim_grid = dim3(ceil(N_h.width / (float) BLOCK_SIZE), ceil(N_h.height / (float) BLOCK_SIZE));
    dim_block = dim3(BLOCK_SIZE, BLOCK_SIZE);



    //KERNEL Launch

    convolution<<<dim_grid, dim_block>>>(N_d, P_d);

    return 0;
}

kernel.cu:ЭТО ГДЕ Я НЕ ЗНАЮ, КАК ДОСТУП К ПОСТОЯННОЙ ПАМЯТИ.

//__constant__ float M_c[FILTER_SIZE][FILTER_SIZE];

__global__ void convolution(Matrix N, Matrix P)
{
    /********************************************************************
    Determine input and output indexes of each thread
    Load a tile of the input image to shared memory
    Apply the filter on the input image tile
    Write the compute values to the output image at the correct indexes
    ********************************************************************/

    //INSERT KERNEL CODE HERE

    //__shared__ float N_shared[BLOCK_SIZE][BLOCK_SIZE];


    //int row = (blockIdx.y * blockDim.y) + threadIdx.y;
    //int col = (blockIdx.x * blockDim.x) + threadIdx.x;

}