10 static __global__ void add(float* ptrDevV1, float* ptrDevV2, int n, float* ptrDevResult);
11 static __device__ float work(float v1, float v2);
13 __global__ void add(float* ptrDevV1, float* ptrDevV2, int n, float* ptrDevResult)
15 const int NB_THREAD = Indice2D::nbThread();
16 const int TID = Indice2D::tid();
22 ptrDevResult[s] = work(ptrDevV1[s], ptrDevV2[s]);
27 __device__ float work(float v1, float v2)
34 cout << "addVectors() ..." << endl;
36 // Inputs (passé en paramètre de la fonction dans un cas général).
37 float v1[] = { 1, 2, 3 };
38 float v2[] = { 10, 20, 30 };
40 // Outputs (renvoyer de la fonction dans un cas général).
43 // Allocation coté GPU.
44 float* ptrDevV1, *ptrDevV2, *ptrDevVResult = 0;
45 const size_t vecSize = 3 * sizeof(float);
46 HANDLE_ERROR(cudaMalloc(&ptrDevV1, vecSize));
47 HANDLE_ERROR(cudaMalloc(&ptrDevV2, vecSize));
48 HANDLE_ERROR(cudaMalloc(&ptrDevVResult, vecSize));
50 HANDLE_ERROR(cudaMemset(ptrDevV1, 0, vecSize));
51 HANDLE_ERROR(cudaMemset(ptrDevV2, 0, vecSize));
52 HANDLE_ERROR(cudaMemset(ptrDevVResult, 0, vecSize));
54 HANDLE_ERROR(cudaMemcpy(ptrDevV1, v1, vecSize, cudaMemcpyHostToDevice));
55 HANDLE_ERROR(cudaMemcpy(ptrDevV2, v2, vecSize, cudaMemcpyHostToDevice));
57 const dim3 dg(2, 2, 1);
58 const dim3 db(2, 2, 1);
59 Device::assertDim(dg, db);
61 add<<<dg, db>>>(ptrDevV1, ptrDevV2, 3, ptrDevVResult);
63 // Barrière implicite de synchronisation ('cudaMemCpy').
64 HANDLE_ERROR(cudaMemcpy(vRes, ptrDevVResult, vecSize, cudaMemcpyDeviceToHost));
66 return vRes[0] == 11 && vRes[1] == 22 && vRes[2] == 33;