X-Git-Url: http://git.euphorik.ch/?a=blobdiff_plain;f=WCudaMSE%2FStudent_Cuda_Image%2Fsrc%2Fcpp%2Fcore%2F04_RayTracing%2Fmoo%2Fhost%2FRayTracing.cu;h=fbbd16073f32e06358914b8ae3ca3a775eb5cd84;hb=1c4b2276e7157acde9a3014b68d5d1667a7d6a44;hp=503459473dcfad67948bc7107f7aa6f5734e642b;hpb=f2c6a4fc79746e2d5c6678699bd2ca93ffc49bcc;p=GPU.git

diff --git a/WCudaMSE/Student_Cuda_Image/src/cpp/core/04_RayTracing/moo/host/RayTracing.cu b/WCudaMSE/Student_Cuda_Image/src/cpp/core/04_RayTracing/moo/host/RayTracing.cu
index 5034594..fbbd160 100644
--- a/WCudaMSE/Student_Cuda_Image/src/cpp/core/04_RayTracing/moo/host/RayTracing.cu
+++ b/WCudaMSE/Student_Cuda_Image/src/cpp/core/04_RayTracing/moo/host/RayTracing.cu
@@ -9,19 +9,44 @@ using namespace std;
 #include "Device.h"
 #include "RayTracingDevice.h"
 #include "Sphere.h"
-
-RayTracing::RayTracing(int w, int h)
-    : w(w), h(h),
-      dg(8, 8, 1),
-      db(32, 32, 1),
-      title("Ray Tracing")
+#include "RayTracingParams.h"
+
+RayTracing::RayTracing(int w, int h, int dg, int db) :
+    t(0), w(w), h(h),
+    dg(dg, dg, 1),
+    db(db, db, 1),
+    alea(42),
+#if CURRENT_MEMORY_MODEL == MEMORY_MODEL_GLOBAL
+    title("Ray Tracing (Global memory)")
+#elif CURRENT_MEMORY_MODEL == MEMORY_MODEL_CONSTANT
+    title("Ray Tracing (Constant memory)")
+#elif CURRENT_MEMORY_MODEL == MEMORY_MODEL_SHARED
+    title("Ray Tracing (Shared memory)")
+#endif
     {
     Device::assertDim(dg, db);
 
-    const int nbSpheres = 10;
-    Sphere* shperes = this->createSpheres(10);
-
-    // Copie les spheres dans la constant memory.
+    Sphere* spheres = this->createSpheres(NB_SPHERES);
+    cout << "sizeof(Sphere): " << sizeof(Sphere) << endl;
+    cout << "Nb of spheres: " << NB_SPHERES << endl;
+    cout << "sizeof(spheres): " << NB_SPHERES * sizeof(Sphere) << endl;
+    cout << "dg: " << this->dg.x << " x " << this->dg.y << endl;
+    cout << "db: " << this->db.x << " x " << this->db.y << endl;
+
+#if CURRENT_MEMORY_MODEL == MEMORY_MODEL_GLOBAL or CURRENT_MEMORY_MODEL == MEMORY_MODEL_SHARED
+    // Copie des spheres dans la global memory.
+    const size_t sizeSpheres = NB_SPHERES * sizeof(Sphere);
+    HANDLE_ERROR(cudaMalloc(&this->ptrDevSpheres, sizeSpheres));
+    HANDLE_ERROR(cudaMemcpy(this->ptrDevSpheres, spheres, sizeSpheres, cudaMemcpyHostToDevice));
+#elif CURRENT_MEMORY_MODEL == MEMORY_MODEL_CONSTANT
+    // Copie des spheres en constant memory.
+    ConstantMemoryLink cmSpheresLink = constantMemorySpheresLink();
+    Sphere* ptrDevConstSperes = (Sphere*)cmSpheresLink.ptrDevTab;
+    size_t sizeALL = cmSpheresLink.sizeAll;
+    HANDLE_ERROR(cudaMemcpy(ptrDevConstSperes, spheres, sizeALL, cudaMemcpyHostToDevice));
+#endif
+
+    cout << "sizeof(Sphere): " << sizeof(Sphere) << endl;
     }
 
 RayTracing::~RayTracing()
@@ -30,14 +55,18 @@ RayTracing::~RayTracing()
 
 void RayTracing::runGPU(uchar4* ptrDevPixels)
     {
-    rayTracing<<<dg,db>>>(ptrDevPixels, this->w, this->h, this->t);
+#if CURRENT_MEMORY_MODEL == MEMORY_MODEL_GLOBAL or CURRENT_MEMORY_MODEL == MEMORY_MODEL_SHARED
+    rayTracing<<<dg, db>>>(ptrDevPixels, this->ptrDevSpheres, this->w, this->h, this->t);
+#elif CURRENT_MEMORY_MODEL == MEMORY_MODEL_CONSTANT
+    rayTracing<<<dg, db>>>(ptrDevPixels, this->w, this->h, this->t);
+#endif
 
-    HANDLE_ERROR(cudaDeviceSynchronize()); // Pour flusher les 'printf' (pour le DEBUG).
+    // HANDLE_ERROR(cudaDeviceSynchronize()); // Pour flusher les 'printf' (pour le DEBUG).
     }
 
 void RayTracing::animationStep()
     {
-    this->t += 0.1; // TODO
+    this->t += 0.005;
     }
 
 int RayTracing::getW()
@@ -63,20 +92,25 @@ string RayTracing::getTitle()
 Sphere* RayTracing::createSpheres(int n)
     {
     Sphere* spheres = new Sphere[n];
-    const float bord = 200;
+    const int RAYON_MIN = 10;
+    const int RAYON_MAX = this->w / 10 - 1;
+
+    const float bord = RAYON_MAX + 1;
 
     for (int i = 0; i < n; ++i)
         {
-        spheres[i].setR(float(this->alea.uniformeAB(20, this->w / 10 - 1)));
+        spheres[i].setR(float(this->alea.uniformeAB(RAYON_MIN, RAYON_MAX)));
 
-        cpu::float3 centre
+        const float3 centre =
             {
-            float(this->alea.uniformeAB(double(bord), double(this->w - bord))),
-            float(this->alea.uniformeAB(double(bord), double(this->h - bord))),
-            float(this->alea.uniformeAB(10.0, 2.0 * this->w))
+            float(this->alea.uniformeAB(double(bord), double(this->w - bord))), // x.
+            float(this->alea.uniformeAB(double(bord), double(this->h - bord))), // y.
+            float(this->alea.uniformeAB(10.0, 2.0 * this->w))                   // z.
             };
 
         spheres[i].setCentre(centre);
         spheres[i].setHueInitial(float(this->alea.uniformeAB(0.0, 1.0)));
         }
+
+    return spheres;
     }