cudaMemcpy struct Gerät zum Host funktioniert nicht

Question

wenn ich Array von Knoten vom Gerät zurück zum Host zu kopieren Ich bekomme Nullen in Node.m [...] anstelle von Werten, obwohl, wenn ich Knoten im Kernel drucken zeigt dies Werte sind richtig eingestellt. Leider kann ich keinen Fehler erkennen, daher bitte ich Sie um Hilfe. Ich kompiliere Code mit Visual Studio Compiler und Rechenkapazität 3. Code von this Antwort funktioniert für mich obwohl.

ich einfügen ganzen Code, aber nur sinnvoll, Teile sind

__global__ void divideLeft(Node * nodes,float * leftSide){...} 

und

divideLeft<<<1,1>>>(dNodes,dLeftSide); 
ERRCHECK(cudaDeviceSynchronize()); 
ERRCHECK(cudaGetLastError()); 
ERRCHECK(cudaMemcpy(nodes,dNodes,sizeof(Node) * heapSize,cudaMemcpyDeviceToHost)); 
printNode(nodes[3]); 

---

#include "cuda_runtime.h" 
#include "device_launch_parameters.h" 

#include <stdio.h> 
#include <conio.h> 
#include <new> 
#include <cmath> 

#define ERRCHECK(ans) { gpuAssert((ans), __FILE__, __LINE__); } 
inline void gpuAssert(cudaError_t code, const char *file, int line, bool abort=true,bool wait=true) 
{ 
    if (code != cudaSuccess) 
    { 
     fprintf(stderr,"GPUassert: %s %s %d\n", cudaGetErrorString(code), file, line); 
     if (wait) getch(); 
     if (abort) exit(code); 
    } 
} 

#define MSIZE 36 
#define INPUT_SIZE(N) N*5 - 3*2 
#define PARENT(i) (i-1)/2 
#define LEFT(i) 2*i + 1 
#define RIGHT(i) 2*i + 2 
#define BOTTOM_HEAP_NODES_COUNT(N) (N-2)/3 //size of input must be 2+3n,n>1 
#define HEAP_SIZE(N) 2*BOTTOM_HEAP_NODES_COUNT(N)-1 
#define FIRST_LEVEL_SIZE 19 
#define ROW_LENGTH 5 
#define FIRST_LVL_MAT_SIZE 5 
#define XY(x,y) x*6+y 

__constant__ int dHigherTreeLevelThreshold; 
__constant__ int dNodesCount; 
__constant__ int dLeftSize; 
__constant__ int dHeapSize; 
__constant__ int dBottomNodes; 
__constant__ int dRemainingNodes; 
__constant__ int dRightCols; 
__constant__ int dInputCount; 

struct Node 
{ 
    float m[MSIZE]; 
    float *x; 
}; 

__device__ __host__ void printNode(Node node); 
__global__ void divideLeft(Node * nodes,float * leftSide) 
{ 
    int idx = blockIdx.x*blockDim.x + threadIdx.x; 
    if(idx>=dBottomNodes) 
     return; 
    int nodeIdx = idx + dRemainingNodes - (idx >= dHigherTreeLevelThreshold)*dBottomNodes; 
// printf("%d %d\n",idx,nodeIdx); 
    Node node = nodes[nodeIdx]; 
    idx*=5*3; 
    node.m[XY(3,3)] = leftSide[idx+2]/3; 
    node.m[XY(3,2)] = leftSide[idx+3]/2; 
    node.m[XY(3,1)] = leftSide[idx+4]; 

    node.m[XY(2,3)] = leftSide[idx+6]/2; 
    node.m[XY(2,2)] = leftSide[idx+7]*2/3; 
    node.m[XY(2,1)] = leftSide[idx+8]; 
    node.m[XY(2,4)] = leftSide[idx+9]; 

    node.m[XY(1,3)] = leftSide[idx+10]; 
    node.m[XY(1,2)] = leftSide[idx+11]; 
    node.m[XY(1,1)] = leftSide[idx+12]; 
    node.m[XY(1,4)] = leftSide[idx+13]; 
    node.m[XY(1,5)] = leftSide[idx+14]; 

    node.m[XY(4,2)] = leftSide[idx+15]; 
    node.m[XY(4,1)] = leftSide[idx+16]; 
    node.m[XY(4,4)] = leftSide[idx+17]*2/3; 
    node.m[XY(4,5)] = leftSide[idx+18]/2; 

    node.m[XY(5,1)] = leftSide[idx+20]; 
    node.m[XY(5,4)] = leftSide[idx+21]/2; 
    node.m[XY(5,5)] = leftSide[idx+22]/3; 
    printNode(node); 
} 

void leftSideInit(float * leftSide,int size) 
{ 
    for(int i = 0;i<size;i++) 
    { 
     leftSide[i] = 1;//(i+1)%26; 
    } 
} 

int main(){ 
    ERRCHECK(cudaSetDevice(0)); 

    int leftCount = 11; 
    int leftSize = leftCount*5; 
    int rightSize = 10; 
    int heapSize = HEAP_SIZE(leftCount); 
    int bottomNodes = BOTTOM_HEAP_NODES_COUNT(leftCount); 
    int greatestPowerOfTwo = pow(2,(int)log2(bottomNodes)); 
    int remainingNodes = heapSize - greatestPowerOfTwo; 

    ERRCHECK(cudaMemcpyToSymbol(dBottomNodes,&bottomNodes,sizeof(int))); 
    ERRCHECK(cudaMemcpyToSymbol(dHigherTreeLevelThreshold,&greatestPowerOfTwo,sizeof(int))); 
    ERRCHECK(cudaMemcpyToSymbol(dRemainingNodes,&remainingNodes,sizeof(int))); 
    ERRCHECK(cudaMemcpyToSymbol(dRightCols,&rightSize,sizeof(int))); 
    ERRCHECK(cudaMemcpyToSymbol(dHeapSize,&heapSize,sizeof(int))); 

    float * leftSide = new float[leftSize]; 
    float * rightSide = new float[rightSize]; 
    Node * nodes = new Node[heapSize]; 
    Node * dNodes = nullptr; 
    float * dLeftSide =nullptr; 
    leftSideInit(leftSide,leftSize); 

    ERRCHECK(cudaMalloc(&dNodes,sizeof(Node)* heapSize)); 
    ERRCHECK(cudaMemset(dNodes,0,sizeof(Node)*heapSize)); 
    ERRCHECK(cudaMalloc(&dLeftSide,leftSize*sizeof(float))); 
    ERRCHECK(cudaMemcpy(dLeftSide,leftSide,leftSize*sizeof(float),cudaMemcpyHostToDevice)); 
    divideLeft<<<1,1>>>(dNodes,dLeftSide); 
    ERRCHECK(cudaDeviceSynchronize()); 
    ERRCHECK(cudaGetLastError()); 
    ERRCHECK(cudaMemcpy(nodes,dNodes,sizeof(Node) * heapSize,cudaMemcpyDeviceToHost)); 
    printNode(nodes[3]); 
    delete [] nodes; 
    cudaFree(dNodes); 

    ERRCHECK(cudaDeviceReset()); 

    getch(); 
    return 0; 
} 

__device__ __host__ void printNode(Node node) 
{ 
    for(int i= 0;i<6;i++) 
     printf("%.3f %.3f %.3f %.3f %.3f %.3f\n",node.m[XY(i,0)],node.m[XY(i,1)],node.m[XY(i,2)],node.m[XY(i,3)],node.m[XY(i,4)],node.m[XY(i,5)]); 

} 

Answer 1

In Ihrem Kernel Sie eine lokale Kopie der Node gemacht, dass Sie arbeiten an:

Node node = nodes[nodeIdx]; 

Der Rest des Kernels fährt fort, Elemente von node, Ihre lokale Kopie zu modifizieren.

Nachdem jedoch alle Änderungen vorgenommen wurden, kopieren Sie die lokale Kopie nie wieder in die globale Kopie, sodass die globale Kopie unverändert bleibt.

Um dies zu beheben, eine Möglichkeit wäre, diese Zeile am Ende Ihres Kernel hinzuzufügen:

nodes[nodeIdx] = node; 

Als Nebenwirkung, stelle ich fest, dass Ihr struct Node eine Zeigervariable enthält:

struct Node 
{ 
    float m[MSIZE]; 
    float *x; 
}; 

Sie sollten sich der Tatsache bewusst sein, dass die Verwendung eines Arrays von Strukturen mit eingebetteten Zeigern einige spezielle Komplexitäten haben kann. Sie verwenden diese Variable (x) noch nicht, also erwähne ich das nur als Kommentar. Sie können sich auf die cuda tag info page für eine kanonische Frage zu diesem Konzept ("Verwenden von Arrays von Zeigern in CUDA") beziehen.