#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 "RayTracingCommon.h"
+
+RayTracing::RayTracing(int w, int h) :
+ t(0), w(w), h(h),
+ dg(8, 8, 1),
+ db(32, 32, 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);
+
+#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()
void RayTracing::runGPU(uchar4* ptrDevPixels)
{
+#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()
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)));
- 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;
}