+++ /dev/null
-#include <iostream>
-
-#include "Indice2D.h"
-#include "cudaTools.h"
-#include "Device.h"
-
-using std::cout;
-using std::endl;
-
-static __global__ void add(float* ptrDevV1, float* ptrDevV2, int n, float* ptrDevResult);
-static __device__ float work(float v1, float v2);
-
-__global__ void add(float* ptrDevV1, float* ptrDevV2, int n, float* ptrDevResult)
- {
- const int NB_THREAD = Indice2D::nbThread();
- const int TID = Indice2D::tid();
-
- int s = TID;
-
- while (s < n)
- {
- ptrDevResult[s] = work(ptrDevV1[s], ptrDevV2[s]);
- s += NB_THREAD;
- }
- }
-
-__device__ float work(float v1, float v2)
- {
- return v1 + v2;
- }
-
-bool addVectors()
- {
- // Inputs (passé en paramètre de la fonction dans un cas général).
- float v1[] = { 1, 2, 3 };
- float v2[] = { 10, 20, 30 };
-
- // Outputs (renvoyer de la fonction dans un cas général).
- float vRes[3];
-
- // Allocation coté GPU.
- float* ptrDevV1, *ptrDevV2, *ptrDevVResult = 0;
- const size_t vecSize = 3 * sizeof(float);
- HANDLE_ERROR(cudaMalloc(&ptrDevV1, vecSize));
- HANDLE_ERROR(cudaMalloc(&ptrDevV2, vecSize));
- HANDLE_ERROR(cudaMalloc(&ptrDevVResult, vecSize));
-
- HANDLE_ERROR(cudaMemset(ptrDevV1, 0, vecSize));
- HANDLE_ERROR(cudaMemset(ptrDevV2, 0, vecSize));
- HANDLE_ERROR(cudaMemset(ptrDevVResult, 0, vecSize));
-
- HANDLE_ERROR(cudaMemcpy(ptrDevV1, v1, vecSize, cudaMemcpyHostToDevice));
- HANDLE_ERROR(cudaMemcpy(ptrDevV2, v2, vecSize, cudaMemcpyHostToDevice));
-
- const dim3 dg(2, 2, 1);
- const dim3 db(2, 2, 1);
- Device::assertDim(dg, db);
-
- add<<<dg, db>>>(ptrDevV1, ptrDevV2, 3, ptrDevVResult);
-
- // Barrière implicite de synchronisation ('cudaMemCpy').
- HANDLE_ERROR(cudaMemcpy(vRes, ptrDevVResult, vecSize, cudaMemcpyDeviceToHost));
-
- return vRes[0] == 11 && vRes[1] == 22 && vRes[2] == 33;
- }