WCudaMSE/Student_Cuda_Image/src/cpp/core/04_RayTracing/moo/host/RayTracing.cu

   1 #include <iostream>
   2 #include <assert.h>
   3 #include <stdio.h>
   4 using namespace std;
   5
   6 #include "AleaTools.h"
   7
   8 #include "RayTracing.h"
   9 #include "Device.h"
  10 #include "RayTracingDevice.h"
  11 #include "Sphere.h"
  12 #include "RayTracingParams.h"
  13
  14 RayTracing::RayTracing(int w, int h, int dg, int db) :
  15     t(0), w(w), h(h),
  16     dg(dg, dg, 1),
  17     db(db, db, 1),
  18     alea(42),
  19 #if CURRENT_MEMORY_MODEL == MEMORY_MODEL_GLOBAL
  20     title("Ray Tracing (Global memory)")
  21 #elif CURRENT_MEMORY_MODEL == MEMORY_MODEL_CONSTANT
  22     title("Ray Tracing (Constant memory)")
  23 #elif CURRENT_MEMORY_MODEL == MEMORY_MODEL_SHARED
  24     title("Ray Tracing (Shared memory)")
  25 #endif
  26     {
  27     Device::assertDim(dg, db);
  28
  29     Sphere* spheres = this->createSpheres(NB_SPHERES);
  30     cout << "sizeof(Sphere): " << sizeof(Sphere) << endl;
  31     cout << "Nb of spheres: " << NB_SPHERES << endl;
  32     cout << "sizeof(spheres): " << NB_SPHERES * sizeof(Sphere) << endl;
  33     cout << "dg: " << this->dg.x << " x " << this->dg.y << endl;
  34     cout << "db: " << this->db.x << " x " << this->db.y << endl;
  35
  36 #if CURRENT_MEMORY_MODEL == MEMORY_MODEL_GLOBAL or CURRENT_MEMORY_MODEL == MEMORY_MODEL_SHARED
  37     // Copie des spheres dans la global memory.
  38     const size_t sizeSpheres = NB_SPHERES * sizeof(Sphere);
  39     HANDLE_ERROR(cudaMalloc(&this->ptrDevSpheres, sizeSpheres));
  40     HANDLE_ERROR(cudaMemcpy(this->ptrDevSpheres, spheres, sizeSpheres, cudaMemcpyHostToDevice));
  41 #elif CURRENT_MEMORY_MODEL == MEMORY_MODEL_CONSTANT
  42     // Copie des spheres en constant memory.
  43     ConstantMemoryLink cmSpheresLink = constantMemorySpheresLink();
  44     Sphere* ptrDevConstSperes = (Sphere*)cmSpheresLink.ptrDevTab;
  45     size_t sizeALL = cmSpheresLink.sizeAll;
  46     HANDLE_ERROR(cudaMemcpy(ptrDevConstSperes, spheres, sizeALL, cudaMemcpyHostToDevice));
  47 #endif
  48
  49     cout << "sizeof(Sphere): " << sizeof(Sphere) << endl;
  50     }
  51
  52 RayTracing::~RayTracing()
  53     {
  54     }
  55
  56 void RayTracing::runGPU(uchar4* ptrDevPixels)
  57     {
  58 #if CURRENT_MEMORY_MODEL == MEMORY_MODEL_GLOBAL or CURRENT_MEMORY_MODEL == MEMORY_MODEL_SHARED
  59     rayTracing<<<dg, db>>>(ptrDevPixels, this->ptrDevSpheres, this->w, this->h, this->t);
  60 #elif CURRENT_MEMORY_MODEL == MEMORY_MODEL_CONSTANT
  61     rayTracing<<<dg, db>>>(ptrDevPixels, this->w, this->h, this->t);
  62 #endif
  63
  64     // HANDLE_ERROR(cudaDeviceSynchronize()); // Pour flusher les 'printf' (pour le DEBUG).
  65     }
  66
  67 void RayTracing::animationStep()
  68     {
  69     this->t += 0.005;
  70     }
  71
  72 int RayTracing::getW()
  73     {
  74     return this->w;
  75     }
  76
  77 int RayTracing::getH()
  78     {
  79     return this->h;
  80     }
  81
  82 float RayTracing::getT()
  83     {
  84     return this->t;
  85     }
  86
  87 string RayTracing::getTitle()
  88     {
  89     return this->title;
  90     }
  91
  92 Sphere* RayTracing::createSpheres(int n)
  93     {
  94     Sphere* spheres = new Sphere[n];
  95     const int RAYON_MIN = 10;
  96     const int RAYON_MAX = this->w / 10 - 1;
  97
  98     const float bord = RAYON_MAX + 1;
  99
 100     for (int i = 0; i < n; ++i)
 101         {
 102         spheres[i].setR(float(this->alea.uniformeAB(RAYON_MIN, RAYON_MAX)));
 103
 104         const float3 centre =
 105             {
 106             float(this->alea.uniformeAB(double(bord), double(this->w - bord))), // x.
 107             float(this->alea.uniformeAB(double(bord), double(this->h - bord))), // y.
 108             float(this->alea.uniformeAB(10.0, 2.0 * this->w))                   // z.
 109             };
 110
 111         spheres[i].setCentre(centre);
 112         spheres[i].setHueInitial(float(this->alea.uniformeAB(0.0, 1.0)));
 113         }
 114
 115     return spheres;
 116     }