+++ /dev/null
-#ifndef RAY_TRACING_COMMON
-#define RAY_TRACING_COMMON
-
-#define NB_SPHERES 2000
-
-
-#define MEMORY_MODEL_GLOBAL 1
-#define MEMORY_MODEL_CONSTANT 2
-#define MEMORY_MODEL_SHARED 3
-
-#define CURRENT_MEMORY_MODEL MEMORY_MODEL_SHARED
-
-#endif
float brightness = 0;
float hue = 0;
- // Sélectionne la sphere la plus proche, il peut ne pas y avoir de sphère.
+ // Sélectionne la sphère la plus proche, il peut ne pas y avoir de sphère trouvée.
for (int i = 0; i < NB_SPHERES; ++i)
{
#if CURRENT_MEMORY_MODEL == MEMORY_MODEL_GLOBAL
#elif CURRENT_MEMORY_MODEL == MEMORY_MODEL_SHARED
const Sphere& sphere = tabSpheresSM[i];
#endif
+
const float hSquare = sphere.hSquare(pixel);
if (sphere.isBelow(hSquare))
}
}
- // Si une sphere a pu être trouvée.
+ // Si une sphère a pu être trouvée.
if (selectedSphereDistance != FLT_MAX)
{
ColorTools::HSB_TO_RVB(hue, 1, brightness, &ptrDevPixels[s]);
#include "ConstantMemoryLink.h"
#include "Sphere.h"
-#include "RayTracingCommon.h"
+#include "RayTracingParams.h"
#if CURRENT_MEMORY_MODEL == MEMORY_MODEL_GLOBAL or CURRENT_MEMORY_MODEL == MEMORY_MODEL_SHARED
class Sphere
{
public:
- /*__host__
- Sphere(float3 centre, float r, float hue) :
- centre(centre),
- hueInitial(hue)
- {
- this->setR(r);
- }*/
-
__host__
__device__
Sphere()
{
}
- /*Sphere(const Sphere& other) :
- r(other.r),
- centre(other.centre),
- rCarre(other.rCarre),
- T(other.T)
- {
- }*/
-
__host__
void setCentre(const float3& centre)
{
}
/**
- * Renvoie la le B de HSB compris entre 0 et 1.
+ * Renvoie la le B de HSB compris entre 0 et 1 dépendant du paramètre 'dz'
+ * lui même compris entre 0 et le rayon 'r'.
*/
__device__
float brightness(float dz) const
private:
float r; // Rayon.
float3 centre; // Position.
-
float rCarre; // Précalul de r².
-
float T; // Utilisé pour l'animation.
};
#endif
#include "Device.h"
#include "RayTracingDevice.h"
#include "Sphere.h"
-#include "RayTracingCommon.h"
+#include "RayTracingParams.h"
-RayTracing::RayTracing(int w, int h) :
+RayTracing::RayTracing(int w, int h, int dg, int db) :
t(0), w(w), h(h),
- dg(8, 8, 1),
- db(32, 32, 1),
+ dg(dg, dg, 1),
+ db(db, db, 1),
alea(42),
#if CURRENT_MEMORY_MODEL == MEMORY_MODEL_GLOBAL
title("Ray Tracing (Global memory)")
Device::assertDim(dg, db);
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.
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);
+ 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);
+ 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()
class RayTracing : public Animable_I
{
public:
- RayTracing(int w, int h);
+ RayTracing(int w, int h, int dg, int db);
~RayTracing();
void runGPU(uchar4* ptrDevPixels) /*override*/;
#include "RayTracingProvider.h"
-RayTracing* RayTracingProvider::create()
+RayTracing* RayTracingProvider::create(int dw, int dh, int dg, int db)
{
- int dw = 16 * 50;
- int dh = 16 * 50;
-
- return new RayTracing(dw, dh);
+ return new RayTracing(dw, dh, dg, db);
}
Image* RayTracingProvider::createGL()
{
+
+ const int dw = 16 * 50;
+ const int dh = 16 * 50;
+ const int dg = 32; // 32x32.
+ const int db = 8; // 8x8.
+
ColorRGB_01* ptrColorTitre = new ColorRGB_01(255, 255, 255);
- return new Image(create(), ptrColorTitre); // both ptr destroy by destructor of ImageFonctionel
+ return new Image(create(dw, dh, dg, db), ptrColorTitre); // both ptr destroy by destructor of ImageFonctionel
}
class RayTracingProvider
{
public:
- static RayTracing* create();
+ static RayTracing* create(int dw, int dh, int dg, int db);
static Image* createGL();
};
#include <iostream>\r
#include <stdlib.h>\r
#include <string.h>\r
+using namespace std;\r
\r
#include "Device.h"\r
#include "cudaTools.h"\r
+#include "Chronos.h"\r
\r
-using std::cout;\r
-using std::endl;\r
-using std::string;\r
+#include "RayTracingProvider.h"\r
+#include "RayTracingParams.h"\r
\r
/*----------------------------------------------------------------------*\\r
|* Declaration *|\r
|* Implementation *|\r
\*---------------------------------------------------------------------*/\r
\r
+void run(uchar4* ptrDevPixels, int dw, int dh, int dg, int db, int nbFrame)\r
+ {\r
+ cout << "Raytracing Without GL ";\r
+#if CURRENT_MEMORY_MODEL == MEMORY_MODEL_GLOBAL\r
+ cout << "(Global memory)" << endl;\r
+#elif CURRENT_MEMORY_MODEL == MEMORY_MODEL_CONSTANT\r
+ cout << "(Constant memory)" << endl;\r
+#elif CURRENT_MEMORY_MODEL == MEMORY_MODEL_SHARED\r
+ cout << "(Shared memory)" << endl;\r
+#endif\r
+\r
+ cout << "Nb frames: " << nbFrame << endl;\r
+\r
+ RayTracing* rayTracing = RayTracingProvider::create(dw, dh, dg, db);\r
+\r
+ Chronos chronos;\r
+ chronos.start();\r
+\r
+ for (int i = 0; i < nbFrame; i++)\r
+ rayTracing->runGPU(ptrDevPixels);\r
+\r
+ HANDLE_ERROR(cudaDeviceSynchronize());\r
+\r
+ chronos.stop();\r
+\r
+ cout << "Time: " << chronos << endl;\r
+ cout << "-------------------" << endl;\r
+ }\r
+\r
/*--------------------------------------*\\r
|* Public *|\r
\*-------------------------------------*/\r
\r
int mainFreeGL(void)\r
{\r
- cout << "\n[FPS] : Free GL, please wait ..." << endl;\r
+ const int NB_FRAME_TO_RENDER = 100;\r
+ const int dw = 16 * 50;\r
+ const int dh = 16 * 50;\r
+\r
+ // dg : 8 -> 128.\r
+ const int dgStart = 8;\r
+ const int dgStop = 128;\r
\r
- rippling0FreeGL(1000); // bad technique\r
+ // db : 8 -> 32\r
+ const int dbStart = 8;\r
+ const int dbStop = 32;\r
+\r
+ uchar4* ptrDevPixels;\r
+ HANDLE_ERROR(cudaMalloc(&ptrDevPixels, dw * dh * sizeof(uchar4)));\r
+\r
+ run(ptrDevPixels, dw, dh, 32, 32, NB_FRAME_TO_RENDER);\r
+ return EXIT_SUCCESS;\r
\r
- // TODO : add other tp here ...\r
+ for (int dg = dgStart; dg <= dgStop; dg *= 2)\r
+ for (int db = dbStart; db <= dbStop; db *= 2)\r
+ run(ptrDevPixels, dw, dh, dg, db, NB_FRAME_TO_RENDER);\r
\r
return EXIT_SUCCESS;\r
}\r
\r
int mainGL(const vector<string>& args)\r
{\r
- const string defaultCommand = "raytracing";\r
+ const string defaultCommand = "demo";\r
const string command = args.size() > 0 ? args[0] : defaultCommand;\r
\r
// AutoViewer<Rippling0Image, Rippling0Provider> rippling0(true, true, 10, 10); // Warmup.\r
\r
if (command == "rippling")\r
- {\r
AutoViewer<Image, RipplingProvider> rippling(true, true, 10, 10);\r
- }\r
else if (command == "mandelbrot")\r
{\r
const bool multiGPU = args.size() >= 2 && args[1] == "--mp";\r
AutoViewer<ImageFonctionel, JuliaProvider> fractalJulia(true, true, 180, 80, false);\r
AutoViewer<ImageFonctionel, NewtonProvider> newtown(true, true, 260, 120, false);\r
AutoViewer<Image, HeatTransfertProvider> heatTransfert(true, false, 1200, 300, false);\r
+ AutoViewer<Image, RayTracingProvider> rayTracing(true, true, 200, 80);\r
GLUTImageViewers::runALL();\r
}\r
else\r
cout << "main" << endl;
if (Device::isCuda())
- {
+ {
GLUTImageViewers::init(argc, argv);
//Device::printAll();
// Server Cuda1: in [0,5]
// Server Cuda2: in [0,2]
- int deviceId = 0;
+ const int deviceId = 5;
+ initCuda(deviceId);
vector<string> args;
for (int i = 1; i < argc; ++i)
if (IS_GL)
{
- return mainGL(args); // Bloquant, Tant qu'une fenetre est ouverte
+ return mainGL(args); // Bloquant, Tant qu'une fenêtre est ouverte.
}
else
{
projects / GPU.git / commitdiff