From f800b4f395cd909b1f0a8aa190d8791bf57ca170 Mon Sep 17 00:00:00 2001 From: Klagarge Date: Sat, 18 Oct 2025 16:37:06 +0200 Subject: [PATCH] feat(warmup): add solutions --- .../00_procedurale/addArrayProcedurale.cu | 38 ++++++++++++++-- .../00_procedurale/useAddArrayProcedurale.cpp | 5 ++- .../device/addArray_device.cu | 19 +++----- .../01_pattern_entrelacement/host/AddArray.cu | 43 +++++++++---------- .../01_pattern_entrelacement/host/AddArray.h | 3 +- .../useAddArrayObject.cpp | 13 ++---- .../01_poo/02_pattern_11/useAddArray11.cpp | 3 +- Student_Cuda_Warmup/src/main/main.cpp | 27 +++++++++--- 8 files changed, 91 insertions(+), 60 deletions(-) diff --git a/Student_Cuda_Warmup/src/core/01_addArray/00_procedurale/addArrayProcedurale.cu b/Student_Cuda_Warmup/src/core/01_addArray/00_procedurale/addArrayProcedurale.cu index 3684138..003c5c0 100755 --- a/Student_Cuda_Warmup/src/core/01_addArray/00_procedurale/addArrayProcedurale.cu +++ b/Student_Cuda_Warmup/src/core/01_addArray/00_procedurale/addArrayProcedurale.cu @@ -9,6 +9,7 @@ #include "cudas.h" #include "GM.h" #include "Kernel.h" +#include "Hardware.h" using std::cout; using std::endl; @@ -32,9 +33,34 @@ static __global__ void kaddArray(float* ptrGMV1 , float* ptrGMV2 , float* ptrGMW * ptrW receptionne le resultat * n nombre de case */ -__host__ bool addArray_procedurale(float* ptrV1 , float* ptrV2 , float* ptrW , int n) // __host__ facultatif +__host__ void addArray_procedurale(float* ptrV1 , float* ptrV2 , float* ptrW , int n) // __host__ facultatif { - // TODO addArray + size_t sizeVector = sizeof(float) * n; + + float* ptrGMV1; + float* ptrGMV2; + float* ptrGMW; + + GM::malloc(&ptrGMV1, sizeVector); + GM::malloc(&ptrGMV2, sizeVector); + GM::malloc(&ptrGMW, sizeVector); + + GM::memcpyHToD(ptrGMV1, ptrV1, sizeVector); + GM::memcpyHToD(ptrGMV2, ptrV2, sizeVector); + + const int MP = Hardware::getMPCount(); + const int CORE_MP = Hardware::getCoreCountMP(); + + dim3 dg(MP, 3, 1); + dim3 db(CORE_MP, 2, 1); // <=1024 + + kaddArray<<>>( ptrGMV1 , ptrGMV2 , ptrGMW , n); + + GM::memcpyDToH(ptrW, ptrGMW, sizeVector); + + GM::free(ptrGMV1); + GM::free(ptrGMV2); + GM::free(ptrGMW); } /*--------------------------------------*\ @@ -50,8 +76,12 @@ __global__ void kaddArray(float* ptrGMV1 , float* ptrGMV2 , float* ptrGMW , int const int TID = Thread2D::tid(); // pattern entrelacement - - // TODO addArray + int s = TID; + while (s < n) + { + ptrGMW[s] = ptrGMV1[s] + ptrGMV2[s]; + s += NB_THREAD; + } } /*----------------------------------------------------------------------*\ diff --git a/Student_Cuda_Warmup/src/core/01_addArray/00_procedurale/useAddArrayProcedurale.cpp b/Student_Cuda_Warmup/src/core/01_addArray/00_procedurale/useAddArrayProcedurale.cpp index 9466103..6fb285a 100755 --- a/Student_Cuda_Warmup/src/core/01_addArray/00_procedurale/useAddArrayProcedurale.cpp +++ b/Student_Cuda_Warmup/src/core/01_addArray/00_procedurale/useAddArrayProcedurale.cpp @@ -37,16 +37,17 @@ bool exemple_addArray_procedurale() ArrayTools::print(ptrV2, n); cout << "--------------------------------------------------------------------" << endl; ArrayTools::print(ptrW, n); + cout<sizeVector = -1; // TODO addArray // octet + db(grid.db) { + this->sizeVector = sizeof(float) * n; - // MM (malloc Device) - { - GM::malloc(&ptrGMV1, sizeVector); - // TODO addArray - } + // MM (malloc Device) + { + GM::malloc(&ptrGMV1, sizeVector); + GM::malloc(&ptrGMV2, sizeVector); + GM::malloc(&ptrGMW, sizeVector); + } } -AddArray::~AddArray() - { +AddArray::~AddArray() { //MM (device free) - { GM::free(ptrGMV1); - // TODO addArray - } - } + GM::free(ptrGMV2); + GM::free(ptrGMW); +} /*--------------------------------------*\ |* Methode *| @@ -61,23 +58,23 @@ AddArray::~AddArray() /** * override */ -void AddArray::run() - { +void AddArray::run() { // MM (copy Host->Device) - { - GM::memcpyHToD(ptrGMV1, ptrV1, sizeVector); - // TODO addArray + { + GM::memcpyHToD(ptrGMV1, ptrV1, sizeVector); + GM::memcpyHToD(ptrGMV2, ptrV2, sizeVector); } - // TODO addArray // call kernel // assynchrone + + addArray<<>>( ptrGMV1 , ptrGMV2 , ptrGMW , n); //Kernel::synchronize();// inutile // MM (Device -> Host) { - // TODO addArray // MM barier de synchronisation implicite + GM::memcpyDToH(ptrW, ptrGMW, sizeVector); } - } +} /*----------------------------------------------------------------------*\ |* End *| diff --git a/Student_Cuda_Warmup/src/core/01_addArray/01_poo/01_pattern_entrelacement/host/AddArray.h b/Student_Cuda_Warmup/src/core/01_addArray/01_poo/01_pattern_entrelacement/host/AddArray.h index 8cb5353..3f39950 100755 --- a/Student_Cuda_Warmup/src/core/01_addArray/01_poo/01_pattern_entrelacement/host/AddArray.h +++ b/Student_Cuda_Warmup/src/core/01_addArray/01_poo/01_pattern_entrelacement/host/AddArray.h @@ -9,8 +9,7 @@ \*---------------------------------------------------------------------*/ -class AddArray - { +class AddArray { /*--------------------------------------*\ |* Constructor *| \*-------------------------------------*/ diff --git a/Student_Cuda_Warmup/src/core/01_addArray/01_poo/01_pattern_entrelacement/useAddArrayObject.cpp b/Student_Cuda_Warmup/src/core/01_addArray/01_poo/01_pattern_entrelacement/useAddArrayObject.cpp index c4849ea..f2ee296 100755 --- a/Student_Cuda_Warmup/src/core/01_addArray/01_poo/01_pattern_entrelacement/useAddArrayObject.cpp +++ b/Student_Cuda_Warmup/src/core/01_addArray/01_poo/01_pattern_entrelacement/useAddArrayObject.cpp @@ -45,12 +45,13 @@ bool exemple_addArray_object() ArrayTools::print(ptrV2, n); cout << "--------------------------------------------------------------------" << endl; ArrayTools::print(ptrW, n); + cout< #include "Couts.h" +#include "Hardware.h" using std::cerr; using std::cout; @@ -29,17 +30,29 @@ static void scalar(bool& isOk); int main(int argc , char** argv) { + // Facultatif + { + Hardware::loadCudaDriverAll(); + Hardware::printAllSimple(); + + const int DEVICE_ID = 0; + Hardware::setDevice(DEVICE_ID); + } + const int IS_VERBOSE = true; - bool isOk = true; + // run + { + bool isOk = true; - // Commenter ce dont vous n'avez pas besoin ci-dessous - scalar(isOk); // commenter dans la methode ci-dessous ce que vous ne voulez pas lancer - array(isOk); // commenter dans la methode ci-dessous ce que vous ne voulez pas lancer + // Commenter ce dont vous n'avez pas besoin ci-dessous + //scalar(isOk); // commenter dans la methode ci-dessous ce que vous ne voulez pas lancer + array(isOk); // commenter dans la methode ci-dessous ce que vous ne voulez pas lancer - Couts::statusln(isOk); + Couts::statusln(isOk); - return isOk ? EXIT_SUCCESS : EXIT_FAILURE; + return isOk ? EXIT_SUCCESS : EXIT_FAILURE; + } } /*-------------------------------------*\ @@ -56,7 +69,7 @@ static void array(bool& isOk) { isOk &= exemple_addArray_procedurale(); isOk &= exemple_addArray_object(); - isOk &= exemple_addArray_11(); + // isOk &= exemple_addArray_11(); } /*----------------------------------------------------------------------*\