CUDA C program that illustrates how to transfer an array of structs to the GPU

copyStructArrayToDevice.cu

/*
  This example should illustrate how to transfer an array of
  structs to the device with CUDA.

  Inspired on https://stackoverflow.com/questions/43175162/copying-array-of-structs-from-host-to-device-cuda
*/

#include <stdio.h>

#define TPB 64  


struct values{
    float x, y, z;
};


values *vals;



__global__ void myKernel(int N, values *d_vals){
    int i = blockIdx.x*blockDim.x + threadIdx.x; 

    if (i>=N) return;

    printf("thread %d, %f %f %f\n", i, d_vals[i].x, d_vals[i].y, d_vals[i].z);
}



void PopulateWithData(int N){
    for (int i = 0; i < N; i++){
        vals[i].x = 1.;
        vals[i].y = 2.;
        vals[i].z = 3.;
    }
}


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

    // handle command-line argument
    if ( argc != 2 ) {
        printf( "usage: %s <size of struct array> \n", argv[0] );
        exit(0);
    }  
    // size of struct array
    sscanf(argv[1], "%d", &N); 

    vals = (values*)malloc(sizeof(values) * N);

    PopulateWithData(N); //populates vals with some data

    values* d_ptr;
    cudaMalloc((void**)&d_ptr, N * sizeof(values));
    cudaMemcpy(d_ptr, vals, N *sizeof(values),cudaMemcpyHostToDevice);

    myKernel<<<(N+TPB-1)/TPB, TPB>>>(N, d_ptr);
    
    cudaFree(d_ptr); 
    free(vals);
}