CUDA如何在主機代碼中聲明常量內存時訪問設備內核中的常量內存？

對於這項記錄，這是作業，所以幫助儘可能少或考慮到這一點。我們使用常量內存來存儲「掩碼矩陣」，該矩陣將用於執行更大矩陣的卷積。當我在主機代碼中時，我使用cudaMemcpyToSymbol（）將掩碼複製到常量內存中。CUDA如何在主機代碼中聲明常量內存時訪問設備內核中的常量內存？

我的問題是，這是複製過來，我啓動我的設備內核代碼如何設備知道在哪裏訪問常量內存掩碼矩陣。在內核啓動時是否需要傳遞一個指針？教授給我們的大部分代碼不應該改變（沒有指向通過的面具的指針），但總是有可能他犯了一個錯誤（儘管這很可能是我對某事的理解）

是否應該將恆定的memeory聲明包含在單獨的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; 

}

來源

2013-10-27 NDEthos

爲什麼你不能只傳遞一個指向常量內存的指針作爲你的卷積函數的參數？ – rasen58

下面

是表示使用__constant__符號的「最小尺寸」的例子。您不需要將任何指針傳遞給__global__函數。

#include <cuda.h> 
#include <cuda_runtime.h> 
#include <stdio.h> 

__constant__ float test_const; 

__global__ void test_kernel(float* d_test_array) { 
    d_test_array[threadIdx.x] = test_const; 
} 

#include <conio.h> 
int main(int argc, char **argv) { 

    float test = 3.f; 

    int N = 16; 

    float* test_array = (float*)malloc(N*sizeof(float)); 

    float* d_test_array; 
    cudaMalloc((void**)&d_test_array,N*sizeof(float)); 

    cudaMemcpyToSymbol(test_const, &test, sizeof(float)); 
    test_kernel<<<1,N>>>(d_test_array); 

    cudaMemcpy(test_array,d_test_array,N*sizeof(float),cudaMemcpyDeviceToHost); 

    for (int i=0; i<N; i++) printf("%i %f\n",i,test_array[i]); 

    getch(); 
    return 0; 
}

來源

2013-10-27 18:31:31 JackOLantern

我可以在一個文件中做到這一點。但是當我有一個kernel.cu和一個main.cu並且它在main中聲明時，我無法在kernel.cu中訪問它。掩碼矩陣是在main中創建的，所以我如何在內核中引用它。當我嘗試說它是未定義的。如果我把__constant__ float M_d [FILTER_SIZE * FILTER_SIZE];在覈心以及我得到它已經被定義的錯誤。 – NDEthos

在「經典」 CUDA編譯你必須定義所有代碼和符號（紋理，常量內存，設備功能），並訪問他們的任何主機API調用（包括內核啓動，結合紋理，複製到符號）同一翻譯單位。這意味着，在同一文件中（或通過同一文件中的多個包含語句）有效。這是因爲「經典」CUDA編譯不包含設備代碼鏈接器。

自CUDA 5發佈以來，有可能使用單獨的編譯模式並將不同的設備代碼對象鏈接到支持它的體系結構中的單個fatbinary載荷。在這種情況下，您需要使用extern關鍵字聲明任何__constant__變量，並且定義該符號恰好一次。

如果你不能使用單獨的編譯，那麼通常的解決方法是在與你的內核相同的.cu文件中定義__constant__符號，幷包含一個小的主機包裝函數，它只調用cudaMemcpyToSymbol來設置__constant__符號題。你可能會對內核調用和紋理操作做同樣的事情。

來源

2013-10-27 20:03:32 talonmies

CUDA如何在主機代碼中聲明常量內存時訪問設備內核中的常量內存？

回答

相關問題