Dynamische Zuweisung von konstantem Speicher in CUDA

Question

Ich versuche, den konstanten Speicher auszunutzen, aber es fällt mir schwer herauszufinden, wie man Arrays verschachtelt. Was ich habe, ist ein Array von Daten, die für interne Daten zählt, aber diese sind für jeden Eintrag unterschiedlich. Basierend auf dem folgenden vereinfachten Code habe ich zwei Probleme. Erstens weiß ich nicht, wie ich die Daten zuordnen soll, auf die die Mitglieder meiner Datenstruktur verweisen. Zweitens, da ich cudaGetSymbolAddress nicht für konstanten Speicher verwenden kann, bin ich nicht sicher, ob ich den globalen Zeiger übergeben kann (was Sie nicht mit einfachem __device__ Speicher tun können).

struct __align(16)__ data{ 
int nFiles; 
int nNames; 
int* files; 
int* names; 
}; 

__device__ __constant__ data *mydata; 

__host__ void initMemory(...) 
{ 
    cudaMalloc((void **) &(mydata), sizeof(data)*dynamicsize); 
    for(int i=; i lessthan dynamicsize; i++) 
    { 
     cudaMemcpyToSymbol(mydata, &(nFiles[i]), sizeof(int), sizeof(data)*i, cudaMemcpyHostToDevice); 
     //... 
     //Problem 1: Allocate & Set mydata[i].files 
    } 
} 

__global__ void myKernel(data *constDataPtr) 
{ 
    //Problem 2: Access constDataPtr[n].files, etc 
} 

int main() 
{ 
    //... 
    myKernel grid, threads (mydata); 
} 

Danke für jede Hilfe angeboten. :-)

Answer 1

Warum verwenden Sie nicht einfach die so genannte "gepackte" Datendarstellung? Mit diesem Ansatz können Sie alle benötigten Daten in eindimensionales Byte-Array platzieren.

[struct data (7*4=28 bytes) 
    [int nFiles=3 (4 bytes)] 
    [int nNames=2 (4 bytes)] 
    [file0 (4 bytes)] 
    [file1 (4 bytes)] 
    [file2 (4 bytes)] 
    [name0 (4 bytes)] 
    [name1 (4 bytes)] 
] 

Answer 2

Ich denke, konstant Speicher 64k ist und Sie können es nicht dynamisch CudaMalloc mit zuteilen: zum Beispiel, wenn Sie

struct data 
{ 
    int nFiles; 
    int nNames; 
    int* files; 
    int* names; 
} 

Sie können diese Daten nur speichern, in dem Array auf diese Weise speichern müssen. Es muss konstant erklärt werden, sagen wir

__device__ __constant__ data mydata[100]; 

Ähnlich brauchen Sie auch nicht, es zu befreien. Außerdem sollten Sie die Referenz nicht über einen Zeiger übergeben, sondern nur als globale Variable darauf zugreifen. Ich habe versucht, eine ähnliche Sache zu tun und es gab mir segfault (in Devicemu).

Answer 3

Diese beiden Themen sollen Ihnen helfen:

http://forums.nvidia.com/index.php?showtopic=30269&hl=embedded

Answer 4

Nein, man kann das tun.

Konstanter Speicher (max. 64 KB) kann vor der Kompilierung nur fest codiert werden.

Sie können jedoch im laufenden Betrieb Texturspeicher zuweisen, der ebenfalls auf dem Gerät zwischengespeichert wird.