tri11paragon
/
COSC-3P93-Project
1
0
Fork 0
Code Issues 1 Pull Requests Packages Projects Releases Wiki Activity

OpenCL works morish

main
Brett 2022-12-12 02:07:59 -05:00
parent 4bfff448de
commit 8f55948344
29 changed files with 807 additions and 678 deletions
Show Stats Download Patch File Download Diff File Expand all files Collapse all files

BIN
Step 3/cmake-build-debug/.ninja_deps
View File

Binary file not shown.

92
Step 3/cmake-build-debug/.ninja_log
View File

@ -1,62 +1,38 @@
# ninja log v5
3726 5450 1669143919820663215 CMakeFiles/Step_3.dir/src/graphics/imgui/imgui_tables.cpp.o 920916506d18de2c
2065 3173 1669143917540598963 CMakeFiles/Step_3.dir/src/graphics/gl/gl.c.o c243981994ce0811
2904 4590 1669143918960638977 CMakeFiles/Step_3.dir/src/graphics/imgui/imgui_demo.cpp.o dc0c8b425019900b
2 1613 1669143915984555117 CMakeFiles/Step_3.dir/src/engine/mpi.cpp.o a59e814306d7978
3 3465 1669143917832607191 CMakeFiles/Step_3.dir/src/engine/world.cpp.o 7d8b0912b2dcd105
3 2904 1669143917272591413 CMakeFiles/Step_3.dir/src/engine/util/debug.cpp.o 41c6f1bdaa5e38f5
1 4373 1669143918740632779 CMakeFiles/Step_3.dir/src/engine/image/image.cpp.o 74ebce9e3746476
1614 4368 1669143918736632665 CMakeFiles/Step_3.dir/src/graphics/debug_gui.cpp.o 738602a72b552e04
3465 4015 1669143918384622746 CMakeFiles/Step_3.dir/src/graphics/imgui/imgui_impl_x11.cpp.o 8002180f0368d0ae
1 2477 1669143916848579462 CMakeFiles/Step_3.dir/src/engine/globals.cpp.o 8fd37dd250c0c57b
2 2065 1669143916436567852 CMakeFiles/Step_3.dir/src/engine/math/colliders.cpp.o 49528379fc85b78
3 2658 1669143917024584425 CMakeFiles/Step_3.dir/src/engine/util/models.cpp.o cb39ce8c856d1d8c
2658 6335 1669143920700688013 CMakeFiles/Step_3.dir/src/graphics/imgui/imgui.cpp.o f512f3e5a828153f
1592 1847 1669144043192138832 Step_3 809bda9fd3dad622
3174 3483 1669143917852607754 CMakeFiles/Step_3.dir/src/graphics/imgui/imgui_impl_glfw.cpp.o 310c2607009eba21
2 3871 1669143918240618691 CMakeFiles/Step_3.dir/src/engine/raytracing.cpp.o afb5968833aed95e
4015 5091 1669143919460653068 CMakeFiles/Step_3.dir/src/graphics/input.cpp.o 572a0a724d80e2d3
3016 5493 1669143919860664340 CMakeFiles/Step_3.dir/src/graphics/imgui/imgui_draw.cpp.o 8b0c755c9ee468e9
3871 6494 1669143920864692632 CMakeFiles/Step_3.dir/src/graphics/imgui/imgui_widgets.cpp.o fe3a1955d79f977a
2477 5452 1669143919820663215 CMakeFiles/Step_3.dir/src/graphics/graphics.cpp.o 8f22bfe01b4af85f
2106 4455 1669143918824635146 CMakeFiles/Step_3.dir/src/graphics/gl/gl.cpp.o 27ebeccd2f38d7f7
3484 3726 1669143918096614633 CMakeFiles/Step_3.dir/src/graphics/imgui/imgui_impl_opengl3.cpp.o f69d69f5fadcde41
2429 4641 1669143919012640444 CMakeFiles/Step_3.dir/src/graphics/gl/shader.cpp.o cdbe2fcc55d335a9
2 3016 1669143917384594567 CMakeFiles/Step_3.dir/src/engine/math/bvh.cpp.o c8d4462e974892bb
2 4046 1669143918412623536 CMakeFiles/Step_3.dir/src/engine/main.cpp.o 9026a4a98b896ce1
1 1592 1669144042940131734 CMakeFiles/Step_3.dir/src/opencl/cl.cpp.o 45b82f6048f98464
3 2106 1669143916476568980 CMakeFiles/Step_3.dir/src/engine/util/parser.cpp.o 80ee4e1261cd645e
3 2428 1669143916796577997 CMakeFiles/Step_3.dir/src/engine/util/loaders.cpp.o c81f0f24f2ebce49
2 1424 1670561165675119419 CMakeFiles/Step_3.dir/src/engine/mpi.cpp.o 5d24537a85ed8416
3 2110 1670561166359138173 CMakeFiles/Step_3.dir/src/engine/util/parser.cpp.o 98940b44ed977a44
2 2124 1670561166375138611 CMakeFiles/Step_3.dir/src/engine/math/colliders.cpp.o e1cb3573834144bf
3 2150 1670561166379138721 CMakeFiles/Step_3.dir/src/engine/util/loaders.cpp.o 7aeb3d1871cf6703
2 2568 1670561166819150786 CMakeFiles/Step_3.dir/src/engine/globals.cpp.o ef21dad2d7d55f95
3 2626 1670561166875152322 CMakeFiles/Step_3.dir/src/engine/util/debug.cpp.o 6a4e11c94bba7102
3 2850 1670561167099158461 CMakeFiles/Step_3.dir/src/engine/util/models.cpp.o a3c3944da9146668
2 3232 1670561167479168881 CMakeFiles/Step_3.dir/src/engine/math/bvh.cpp.o bfaa9112c6c83140
3039 4673 1670818921470590191 CMakeFiles/Step_3.dir/src/graphics/imgui/imgui_tables.cpp.o 98cb1b8b0c737d0d
2111 3258 1670561167507169650 CMakeFiles/Step_3.dir/src/graphics/gl/gl.c.o 1829c9b0491e17e
3 3345 1670561167595172063 CMakeFiles/Step_3.dir/src/engine/world.cpp.o c64baeb5ba601bb4
2501 4061 1670818920858574640 CMakeFiles/Step_3.dir/src/graphics/imgui/imgui_demo.cpp.o 154c013a52483b59
2 1393 1670818918194506956 CMakeFiles/Step_3.dir/src/engine/mpi.cpp.o 5d24537a85ed8416
2 2070 1670825558196061243 CMakeFiles/Step_3.dir/src/engine/world.cpp.o c64baeb5ba601bb4
3 2501 1670818919298535005 CMakeFiles/Step_3.dir/src/engine/util/debug.cpp.o 6a4e11c94bba7102
2 3755 1670818920550566815 CMakeFiles/Step_3.dir/src/engine/image/image.cpp.o 27aa2206b4129a91
1393 3888 1670818920686570270 CMakeFiles/Step_3.dir/src/graphics/debug_gui.cpp.o db185e4c05df35c6
2559 3039 1670818919838548725 CMakeFiles/Step_3.dir/src/graphics/imgui/imgui_impl_x11.cpp.o affde06b41e3251b
2 2138 1670825558264063094 CMakeFiles/Step_3.dir/src/opencl/open_ray_tracing.cpp.o 1e13ffe3688667ce
2 2268 1670818919066529111 CMakeFiles/Step_3.dir/src/engine/globals.cpp.o ef21dad2d7d55f95
2 2027 1670818918826523013 CMakeFiles/Step_3.dir/src/engine/math/colliders.cpp.o e1cb3573834144bf
3 2559 1670818919358536530 CMakeFiles/Step_3.dir/src/engine/util/models.cpp.o a3c3944da9146668
2268 5802 1670818922598618848 CMakeFiles/Step_3.dir/src/graphics/imgui/imgui.cpp.o 4939d3e79602c127
1237 1467 1670825715064323231 Step_3 cf178136bfdaf811
3258 3505 1670561167755176449 CMakeFiles/Step_3.dir/src/graphics/imgui/imgui_impl_glfw.cpp.o 3708a9542084698e
2 3528 1670561167775176997 CMakeFiles/Step_3.dir/src/engine/raytracing.cpp.o f1ea987e58fdb6e
2 2500 1670825558628073004 CMakeFiles/Step_3.dir/src/engine/raytracing.cpp.o f1ea987e58fdb6e
3285 4389 1670818921190583075 CMakeFiles/Step_3.dir/src/graphics/input.cpp.o 78c22cd6c124c15f
2557 4753 1670818921550592224 CMakeFiles/Step_3.dir/src/graphics/imgui/imgui_draw.cpp.o 5fdd1c1c2474d086
2 2306 1670825558432067666 CMakeFiles/Step_3.dir/src/graphics/graphics.cpp.o 3e9ee255e3b079b6
3065 5675 1670818922474615699 CMakeFiles/Step_3.dir/src/graphics/imgui/imgui_widgets.cpp.o d66bb60c1727d1ac
1623 3745 1670818920542566612 CMakeFiles/Step_3.dir/src/graphics/gl/gl.cpp.o 60ee6de1376d9db2
3505 3710 1670561167959182040 CMakeFiles/Step_3.dir/src/graphics/imgui/imgui_impl_opengl3.cpp.o 11f263dd8d9ae8a1
3345 3864 1670561168115186318 CMakeFiles/Step_3.dir/src/graphics/imgui/imgui_impl_x11.cpp.o affde06b41e3251b
2 4074 1670561168319191913 CMakeFiles/Step_3.dir/src/engine/image/image.cpp.o 27aa2206b4129a91
2 4080 1670561168327192132 CMakeFiles/Step_3.dir/src/engine/main.cpp.o 9aae8e577719f01
1424 4249 1670561168499196848 CMakeFiles/Step_3.dir/src/graphics/debug_gui.cpp.o db185e4c05df35c6
2124 4486 1670561168735203318 CMakeFiles/Step_3.dir/src/graphics/gl/gl.cpp.o 60ee6de1376d9db2
2150 4490 1670561168739203427 CMakeFiles/Step_3.dir/src/graphics/gl/shader.cpp.o 2a1ad2717bf993ad
2850 4510 1670561168759203978 CMakeFiles/Step_3.dir/src/graphics/imgui/imgui_demo.cpp.o 154c013a52483b59
3864 5094 1670561169347220099 CMakeFiles/Step_3.dir/src/graphics/input.cpp.o 78c22cd6c124c15f
2568 5492 1670561169739230847 CMakeFiles/Step_3.dir/src/graphics/graphics.cpp.o 3e9ee255e3b079b6
3528 5557 1670561169807232711 CMakeFiles/Step_3.dir/src/graphics/imgui/imgui_tables.cpp.o 98cb1b8b0c737d0d
3233 5811 1670561170059239620 CMakeFiles/Step_3.dir/src/graphics/imgui/imgui_draw.cpp.o 5fdd1c1c2474d086
4074 6159 1670561170407249163 CMakeFiles/Step_3.dir/src/opencl/cl.cpp.o a2b56fe5c1c9fba8
4081 6449 1670561170699257167 CMakeFiles/Step_3.dir/src/opencl/open_ray_tracing.cpp.o 1e13ffe3688667ce
2627 6575 1670561170819260459 CMakeFiles/Step_3.dir/src/graphics/imgui/imgui.cpp.o 4939d3e79602c127
3710 6749 1670561170995265284 CMakeFiles/Step_3.dir/src/graphics/imgui/imgui_widgets.cpp.o d66bb60c1727d1ac
6749 7044 1670561171291273400 Step_3 cf178136bfdaf811
2 1863 1670561237929087985 CMakeFiles/Step_3.dir/src/opencl/open_ray_tracing.cpp.o 1e13ffe3688667ce
1863 2127 1670561238193095128 Step_3 cf178136bfdaf811
2 1862 1670561278262179904 CMakeFiles/Step_3.dir/src/opencl/open_ray_tracing.cpp.o 1e13ffe3688667ce
1862 2142 1670561278538187380 Step_3 cf178136bfdaf811
1975 3990 1670818920790572913 CMakeFiles/Step_3.dir/src/graphics/gl/shader.cpp.o 2a1ad2717bf993ad
2 2557 1670818919354536428 CMakeFiles/Step_3.dir/src/engine/math/bvh.cpp.o bfaa9112c6c83140
1 2633 1670825558760076597 CMakeFiles/Step_3.dir/src/engine/main.cpp.o 9aae8e577719f01
3745 5388 1670818922186608382 CMakeFiles/Step_3.dir/src/opencl/cl.cpp.o a2b56fe5c1c9fba8
3 1623 1670818918422512748 CMakeFiles/Step_3.dir/src/engine/util/parser.cpp.o 98940b44ed977a44
1 1237 1670825714832316941 CMakeFiles/Step_3.dir/src/engine/util/loaders.cpp.o 7aeb3d1871cf6703
2 1224 1670825725604609094 CMakeFiles/Step_3.dir/src/engine/util/loaders.cpp.o 7aeb3d1871cf6703
1224 1432 1670825725812614734 Step_3 cf178136bfdaf811
1 1288 1670825808594857943 CMakeFiles/Step_3.dir/src/engine/util/loaders.cpp.o 7aeb3d1871cf6703
1288 1515 1670825808818864011 Step_3 cf178136bfdaf811
1 1258 1670825836395610588 CMakeFiles/Step_3.dir/src/engine/util/loaders.cpp.o 7aeb3d1871cf6703
1258 1486 1670825836619616654 Step_3 cf178136bfdaf811
1 1253 1670825921749919426 CMakeFiles/Step_3.dir/src/engine/util/loaders.cpp.o 7aeb3d1871cf6703
1253 1464 1670825921957925050 Step_3 cf178136bfdaf811

BIN
Step 3/cmake-build-debug/Step_3
View File

Binary file not shown.

4
Step 3/cmake-build-debug/Testing/Temporary/LastTest.log
View File

@ -1,3 +1,3 @@
Start testing: Dec 09 00:54 EST
Start testing: Dec 12 01:18 EST
----------------------------------------------------------
End testing: Dec 09 00:54 EST
End testing: Dec 12 01:18 EST

BIN
Step 3/cmake-build-release-mpi/.ninja_deps
View File

Binary file not shown.

10
Step 3/cmake-build-release-mpi/.ninja_log
View File

@ -13,3 +13,13 @@
959 2684 1669735767860797367 CMakeFiles/Step_3.dir/src/opencl/cl.cpp.o 7e15a029d7ca916c
2 2038 1669735767212779629 CMakeFiles/Step_3.dir/src/engine/util/parser.cpp.o a8f47b5a2bce1b2a
2 2041 1669735767216779738 CMakeFiles/Step_3.dir/src/engine/util/loaders.cpp.o 971186349d935f91
1 961 1670775820121593228 CMakeFiles/Step_3.dir/src/engine/globals.cpp.o ac7c6dc8b9dc3fd1
2 1307 1670775820469601648 CMakeFiles/Step_3.dir/src/engine/math/colliders.cpp.o 8a7658a9bb94b79
2 1551 1670775820713607554 CMakeFiles/Step_3.dir/src/engine/util/debug.cpp.o aec80b8424e96bfd
2 1789 1670775820949613264 CMakeFiles/Step_3.dir/src/engine/math/bvh.cpp.o 73e310373b3e9e3d
2 1790 1670775820949613264 CMakeFiles/Step_3.dir/src/engine/util/models.cpp.o 781ccf05f86a9476
3 1867 1670775821029615201 CMakeFiles/Step_3.dir/src/engine/util/parser.cpp.o 49df7154b35ebe07
3 2034 1670775821197619266 CMakeFiles/Step_3.dir/src/engine/world.cpp.o e08a39ed946f4838
2 2167 1670775821329622462 CMakeFiles/Step_3.dir/src/engine/util/loaders.cpp.o f94d15a8dacd0fa7
961 2807 1670775821969637949 CMakeFiles/Step_3.dir/src/opencl/cl.cpp.o c1eccdb992e8345f
2 7810 1670775826969758951 CMakeFiles/Step_3.dir/src/engine/image/image.cpp.o 1ef1b77523471449

BIN
Step 3/cmake-build-release-mpi/CMakeFiles/Step_3.dir/src/engine/globals.cpp.o
View File

Binary file not shown.

BIN
Step 3/cmake-build-release-mpi/CMakeFiles/Step_3.dir/src/engine/image/image.cpp.o
View File

Binary file not shown.

BIN
Step 3/cmake-build-release-mpi/CMakeFiles/Step_3.dir/src/engine/math/bvh.cpp.o
View File

Binary file not shown.

BIN
Step 3/cmake-build-release-mpi/CMakeFiles/Step_3.dir/src/engine/math/colliders.cpp.o
View File

Binary file not shown.

BIN
Step 3/cmake-build-release-mpi/CMakeFiles/Step_3.dir/src/engine/util/debug.cpp.o
View File

Binary file not shown.

BIN
Step 3/cmake-build-release-mpi/CMakeFiles/Step_3.dir/src/engine/util/loaders.cpp.o
View File

Binary file not shown.

BIN
Step 3/cmake-build-release-mpi/CMakeFiles/Step_3.dir/src/engine/util/models.cpp.o
View File

Binary file not shown.

BIN
Step 3/cmake-build-release-mpi/CMakeFiles/Step_3.dir/src/engine/util/parser.cpp.o
View File

Binary file not shown.

BIN
Step 3/cmake-build-release-mpi/CMakeFiles/Step_3.dir/src/engine/world.cpp.o
View File

Binary file not shown.

BIN
Step 3/cmake-build-release-mpi/CMakeFiles/Step_3.dir/src/opencl/cl.cpp.o
View File

Binary file not shown.

4
Step 3/cmake-build-release-mpi/Testing/Temporary/LastTest.log
View File

@ -1,3 +1,3 @@
Start testing: Dec 06 14:15 EST
Start testing: Dec 11 11:12 EST
----------------------------------------------------------
End testing: Dec 06 14:15 EST
End testing: Dec 11 11:12 EST

39
Step 3/include/engine/raytracing.h
View File

@ -44,7 +44,8 @@ namespace Raytracing {
Vec4 up{0, 1, 0};
public:
Camera(PRECISION_TYPE fov, const Image& image): image(image),
Camera(PRECISION_TYPE fov, const Image& image):
image(image),
aspectRatio(double(image.getWidth()) / double(image.getHeight())) {
// scale the viewport height based on the camera's FOV
tanFovHalf = tan(degreeeToRadian(fov) / 2);
@ -72,8 +73,6 @@ namespace Raytracing {
void setPosition(const Vec4& pos) { this->position = pos; }
void setRotation(PRECISION_TYPE yaw, PRECISION_TYPE pitch);
// the follow utility functions are actually taking forever to get right
// I can't tell if my projection calculation is off or the view calc?
// got to install GLM to test which function works and which does. Maybe they are both bad. or Maybe it's my matrix impl
@ -96,37 +95,6 @@ namespace Raytracing {
//return Mat4x4{projectG};
}
[[nodiscard]] Mat4x4 view(const Vec4& lookAtPos) const {
Mat4x4 view;
auto w = (position - lookAtPos).normalize(); // forward
auto u = (Vec4::cross(up, w)).normalize(); // right
auto v = Vec4::cross(w, u); // up
view.m00(float(w.x()));
view.m01(float(w.y()));
view.m02(float(w.z()));
view.m03(float(w.w()));
view.m10(float(u.x()));
view.m11(float(u.y()));
view.m12(float(u.z()));
view.m13(float(u.w()));
view.m20(float(v.x()));
view.m21(float(v.y()));
view.m22(float(v.z()));
view.m23(float(v.w()));
// view matrix are inverted, dot product to simulate translate matrix multiplication
view.m30(-float(Vec4::dot(u, position)));
view.m31(-float(Vec4::dot(v, position)));
view.m32(-float(Vec4::dot(w, position)));
view.m33(1);
return view;
}
Mat4x4 view(PRECISION_TYPE yaw, PRECISION_TYPE pitch);
[[nodiscard]] inline Vec4 getPosition() const { return position; };
@ -205,7 +173,8 @@ namespace Raytracing {
// likely due to not over generating unit vectors biased towards the corners
}
RayCaster(Camera& c, Image& i, World& world, const Parser& p): camera(c), image(i), world(world) {
RayCaster(Camera& c, Image& i, World& world, const Parser& p):
camera(c), image(i), world(world) {
world.generateBVH();
}

7
Step 3/include/engine/util/memory_util.h
View File

@ -43,7 +43,8 @@ namespace Raytracing {
auto x = getBytes((float) vec.x());
auto y = getBytes((float) vec.y());
auto z = getBytes((float) vec.z());
auto w = getBytes((float) vec.w());
// we don't really use the w vector but this is a TODO
//auto w = getBytes((float) vec.w());
// write the bytes as a packed vector.
for (auto b: x)
array[offset++] = b;
@ -51,8 +52,8 @@ namespace Raytracing {
array[offset++] = b;
for (auto b: z)
array[offset++] = b;
for (auto b: w)
array[offset++] = b;
//for (auto b: w)
// array[offset++] = b;
}
/**

4
Step 3/include/graphics/graphics.h
View File

@ -150,8 +150,10 @@ namespace Raytracing {
RayCaster& mRaycaster,
Parser& mParser,
Camera& mCamera)
: m_window(mWindow), m_mainImage(mMainImage), m_imageShader(mImageShader), m_worldShader(mWorldShader), m_raycaster(mRaycaster),
:
m_window(mWindow), m_mainImage(mMainImage), m_imageShader(mImageShader), m_worldShader(mWorldShader), m_raycaster(mRaycaster),
m_parser(mParser), m_camera(mCamera), m_world(world) {}
std::pair<Mat4x4, Mat4x4> getCameraMatrices();
void draw();
};
}

2
Step 3/include/opencl/open_ray_tracing.h
View File

@ -27,6 +27,7 @@ namespace Raytracing {
private:
CLProgram* program;
Image& image;
Camera& camera;
size_t localWorks[2]{8, 8};
size_t maxTriangleSize = 0;
size_t objectCount = 0;
@ -36,6 +37,7 @@ namespace Raytracing {
~OpenClRaytracer();
void storeObjects(unsigned char* buffer, size_t totalWorldBytes);
void updateCameraInformation();
unsigned char* createObjectBuffer(const std::vector<Object*>& objects, size_t totalWorldBytes);

75
Step 3/resources/opencl/randoms_git.cl Normal file
View File

@ -0,0 +1,75 @@
/**
@file
Implements a 64-bit Permutated Congruential generator (PCG-XSH-RR).
M. E. O’Neill, Pcg: A family of simple fast space-efficient statistically good algorithms for random number generation, ACM Transactions on Mathematical Software.
FROM: https://github.com/bstatcomp/RandomCL/blob/master/generators/pcg6432.cl
*/
#define RNG32
#define PCG6432_FLOAT_MULTI 2.3283064365386963e-10f
#define PCG6432_DOUBLE2_MULTI 2.3283064365386963e-10
#define PCG6432_DOUBLE_MULTI 5.4210108624275221700372640e-20
/**
State of pcg6432 RNG.
*/
typedef unsigned long pcg6432_state;
#define PCG6432_XORSHIFTED(s) ((uint)((((s) >> 18u) ^ (s)) >> 27u))
#define PCG6432_ROT(s) ((s) >> 59u)
#define pcg6432_macro_uint(state) ( \
state = state * 6364136223846793005UL + 0xda3e39cb94b95bdbUL, \
(PCG6432_XORSHIFTED(state) >> PCG6432_ROT(state)) | (PCG6432_XORSHIFTED(state) << ((-PCG6432_ROT(state)) & 31)) \
)
/**
Generates a random 32-bit unsigned integer using pcg6432 RNG.
@param state State of the RNG to use.
*/
#define pcg6432_uint(state) _pcg6432_uint(&state)
unsigned int _pcg6432_uint(pcg6432_state* state){
ulong oldstate = *state;
*state = oldstate * 6364136223846793005UL + 0xda3e39cb94b95bdbUL;
uint xorshifted = ((oldstate >> 18u) ^ oldstate) >> 27u;
uint rot = oldstate >> 59u;
return (xorshifted >> rot) | (xorshifted << ((-rot) & 31));
}
/**
Seeds pcg6432 RNG.
@param state Variable, that holds state of the generator to be seeded.
@param seed Value used for seeding. Should be randomly generated for each instance of generator (thread).
*/
void pcg6432_seed(pcg6432_state* state, unsigned long j){
*state=j;
}
/**
Generates a random 64-bit unsigned integer using pcg6432 RNG.
@param state State of the RNG to use.
*/
#define pcg6432_ulong(state) ((((ulong)pcg6432_uint(state)) << 32) | pcg6432_uint(state))
/**
Generates a random float using pcg6432 RNG.
@param state State of the RNG to use.
*/
#define pcg6432_float(state) (pcg6432_uint(state)*PCG6432_FLOAT_MULTI)
/**
Generates a random double using pcg6432 RNG.
@param state State of the RNG to use.
*/
#define pcg6432_double(state) (pcg6432_ulong(state)*PCG6432_DOUBLE_MULTI)
/**
Generates a random double using pcg6432 RNG. Generated using only 32 random bits.
@param state State of the RNG to use.
*/
#define pcg6432_double2(state) (pcg6432_uint(state)*PCG6432_DOUBLE2_MULTI)
/*
END OF SOURCED RANDOM GENERATOR
*/

33
Step 3/resources/opencl/raytracer.cl
View File

@ -37,8 +37,8 @@ struct Ray {
float4 inverseDirection;
};
float4 along(struct Ray ray, float length) {
return ray.start + length * ray.direction;
float4 along(struct Ray* ray, float length) {
return ray->start + length * ray->direction;
}
float magnitude(float4 vec){
@ -64,7 +64,7 @@ struct ScatterResults {
};
// as close to a 1 to 1 copy from the cpu ray tracer
bool checkForTriangleIntersection(struct HitData* data, struct Triangle theTriangle, float4 position, struct Ray ray, float min, float max) {
bool checkForTriangleIntersection(struct HitData* data, struct Triangle theTriangle, float4 position, struct Ray* ray, float min, float max) {
const float EPSILON = 0.0000001f;
float4 edge1, edge2, s, h, q;
@ -72,7 +72,7 @@ bool checkForTriangleIntersection(struct HitData* data, struct Triangle theTrian
edge1 = (theTriangle.vertex2 + position) - (theTriangle.vertex1 + position);
edge2 = (theTriangle.vertex3 + position) - (theTriangle.vertex1 + position);
h = cross(ray.direction, edge2);
h = cross(ray->direction, edge2);
a = dot(edge1, h);
if (a > -EPSILON && a < EPSILON) {
@ -80,7 +80,7 @@ bool checkForTriangleIntersection(struct HitData* data, struct Triangle theTrian
}
f = 1.0f / a;
s = ray.start - (theTriangle.vertex1 + position);
s = ray->start - (theTriangle.vertex1 + position);
u = f * dot(s, h);
if (u < 0.0f || u > 1.0f) {
@ -88,7 +88,7 @@ bool checkForTriangleIntersection(struct HitData* data, struct Triangle theTrian
}
q = cross(s, edge1);
v = f * dot(ray.direction, q);
v = f * dot(ray->direction, q);
if (v < 0.0f || u + v > 1.0f) {
return false;
}
@ -155,19 +155,18 @@ float4 getVector(__global unsigned char* buffer, unsigned long* currentByte){
return val;
}
struct Ray projectRay(float x, float y){
struct Ray* projectRay(struct Ray* ray, float x, float y){
float transformedX = (x / (imageWidth - 1));
float transformedY = (y / (imageHeight - 1));
struct Ray ray;
ray.start = cameraPosition;
ray.direction = imageOrigin + transformedX * horizontalAxis + transformedY * verticalAxis - cameraPosition;
ray.inverseDirection = 1.0f / ray.direction;
ray->start = cameraPosition;
ray->direction = imageOrigin + transformedX * horizontalAxis + transformedY * verticalAxis - cameraPosition;
ray->inverseDirection = 1.0f / ray->direction;
return ray;
}
bool checkForWorldIntersection(struct HitData* data, __global struct Object* objects, struct Ray ray, float min, float max){
bool checkForWorldIntersection(struct HitData* data, __global struct Object* objects, struct Ray* ray, float min, float max){
data->length = max;
bool hit = false;
// brute-force check on all the objects in the world.
@ -182,7 +181,7 @@ bool checkForWorldIntersection(struct HitData* data, __global struct Object* obj
return hit;
}
bool scatter(struct ScatterResults* results, __global unsigned char* randoms, struct Ray ray, struct HitData data){
bool scatter(struct ScatterResults* results, __global unsigned char* randoms, struct Ray* ray, struct HitData data){
const float EPSILON = 0.0000001f;
int x = get_global_id(0);
unsigned long cb = x * sizeof(float4);
@ -201,8 +200,7 @@ bool scatter(struct ScatterResults* results, __global unsigned char* randoms, st
return true;
}
float4 raycastI(__global unsigned char* randoms, __global struct Object* objects, struct Ray ray){
struct Ray localRay = ray;
float4 raycastI(__global unsigned char* randoms, __global struct Object* objects, struct Ray* ray){
float4 color = (float4)(1.0f, 1.0f, 1.0f, 1.0f);
for (int i = 0; i < MAX_DEPTH; i++){
struct HitData hit;
@ -236,9 +234,12 @@ __kernel void raycast(__write_only image2d_t outputImage, __global struct Object
float4 randomVector = getVector(randoms, &cb);
int y = get_global_id(1);
struct Ray casterRay;
projectRay(&casterRay, x + randomVector.x, y + randomVector.z);
float4 color = (float4)(0.0);
//for (int i = 0; i < MAX_PER_PIXEL; i++){
color = color + raycastI(randoms, objects, projectRay(x + randomVector.x, y + randomVector.z));
color = color + raycastI(randoms, objects, &casterRay);
//}
float scaleFactor = 1.0 / MAX_PER_PIXEL;
write_imagef(outputImage, (int2)(x, y), (float4)(sqrt(color.x * scaleFactor), sqrt(color.y * scaleFactor), sqrt(color.z * scaleFactor), 1.0f));

170
Step 3/resources/opencl/sphereray.cl
View File

@ -3,78 +3,76 @@
#define objectCount 6
#define imageWidth 800
#define imageHeight 600
#define imageOrigin (float4)(0,0,0,0)
#define horizontalAxis (float4)(0,0,0,0)
#define verticalAxis (float4)(0,0,0,0)
#define cameraPosition (float4)(0,0,0,0)
/*#define imageOrigin (float3)(0,0,0)
#define horizontalAxis (float3)(0,0,0)
#define verticalAxis (float3)(0,0,0)
#define cameraPosition (float3)(0,0,0) */
#define MAX_DEPTH 500
#define MAX_PER_PIXEL 500
#define MAX_DEPTH 50
#define MAX_PER_PIXEL 50
#define PI 3.1415926535897
#include "randoms_git.cl"
struct Ray {
// the starting point for our ray
float4 start;
float3 start;
// and the direction it is currently traveling
float4 direction;
float4 inverseDirection;
float3 direction;
float3 inverseDirection;
};
struct HitData {
// the hit point on the object
float4 hitPoint;
float3 hitPoint;
// the normal of that hit point
float4 normal;
float3 normal;
// the length of the vector from its origin in its direction.
float length;
// Texture UV Coords.
float u, v;
};
struct ScatterResults {
// returns true to recast the ray with the provided ray
bool scattered;
// the new ray to be cast if scattered
struct Ray newRay;
// the color of the material
float4 attenuationColor;
// required because for some reason OpenCL stores vectors in a really weird byte order??
// this prevents all of the graphical issues + allows us to assume the order *no platform dependance*
struct Vec {
float x, y, z;
};
float4 along(struct Ray ray, float length) {
float3 randomVector(unsigned long seed){
return ((float3)(pcg6432_float(seed), pcg6432_float(seed), pcg6432_float(seed)) * 2) - 1;
}
float3 along(struct Ray ray, float length) {
return ray.start + length * ray.direction;
}
float lengthSquared(float4 vec){
return vec.x * vec.x + vec.y * vec.y + vec.z * vec.z + vec.w * vec.w;
float lengthSquared(float3 vec){
return vec.x * vec.x + vec.y * vec.y + vec.z * vec.z;
}
float magnitude(float4 vec){
float magnitude(float3 vec){
return sqrt(lengthSquared(vec));
}
// // // reads one unsigned long of data from the objects buffer.
float4 getVector(__global unsigned char* buffer, unsigned long* currentByte){
float4 val = *((global float4*)(buffer));
*currentByte += sizeof(float4);
return val;
}
struct Ray projectRay(float x, float y){
struct Ray projectRay(__global struct Vec* cameraData, float x, float y){
float transformedX = (x / (imageWidth - 1));
float transformedY = (y / (imageHeight - 1));
float3 cameraPosition = (float3)(cameraData[0].x, cameraData[0].y, cameraData[0].z);
float3 verticalAxis = (float3)(cameraData[1].x, cameraData[1].y, cameraData[1].z);
float3 horizontalAxis = (float3)(cameraData[2].x, cameraData[2].y, cameraData[2].z);
float3 imageOrigin = (float3)(cameraData[3].x, cameraData[3].y, cameraData[3].z);
struct Ray ray;
ray.start = cameraPosition;
ray.direction = imageOrigin + transformedX * horizontalAxis + transformedY * verticalAxis - cameraPosition;
ray.direction = (imageOrigin + (transformedX * horizontalAxis) + (transformedY * verticalAxis)) - cameraPosition;
ray.inverseDirection = 1.0f / ray.direction;
return ray;
}
bool checkIfHit(struct HitData* data, struct Ray ray, float4 position, float radius, float min, float max){
bool checkIfHit(struct HitData* data, struct Ray ray, float3 position, float radius, float min, float max){
float radiusSquared = radius * radius;
float4 rayWRTSphere = ray.start - position;
float3 rayWRTSphere = ray.start - position;
// now determine the discriminant for the quadratic formula for the function of line sphere intercept
float a = lengthSquared(ray.direction);
float b = dot(rayWRTSphere, ray.direction);
@ -101,39 +99,107 @@ bool checkIfHit(struct HitData* data, struct Ray ray, float4 position, float rad
}
}
// the hit point is where the ray is when extended to the root
float4 RayAtRoot = along(ray,root);
float3 RayAtRoot = along(ray,root);
// The normal of a sphere is just the point of the hit minus the center position
float4 normal = (RayAtRoot - position) / radius;
float3 normal = (RayAtRoot - position) / radius;
// calculate the uv coords and normalize to [0, 1]
float u = (atan2(-normal.z, normal.x) + PI) / (2 * PI);
float v = acos(normal.y) / PI;
// have to invert the v since we have to invert the v again later due to triangles
data->hitPoint = RayAtRoot;
data->normal = normal;
data->length = root;
data->u = u;
data->v = 1.0 - v;
return true;
}
bool scatter(struct Ray* ray, struct HitData data, int currentDepth){
const float EPSILON = 0.0000001f;
int x = get_global_id(0);
int y = get_global_id(0);
pcg6432_state state;
unsigned long seed = x * y * currentDepth;
pcg6432_seed(&state, seed);
float3 newRay = data.normal + normalize(randomVector(state));
// rays that are close to zero are liable to floating point precision errors
if (newRay.x < EPSILON && newRay.y < EPSILON && newRay.z < EPSILON)
newRay = data.normal;
__kernel void raycast(__write_only image2d_t outputImage, __global unsigned char* objects, __global unsigned char* randoms) {
ray->start = data.hitPoint;
ray->direction = newRay;
ray->inverseDirection = 1/ray->direction;
return true;
}
int checkWorldForIntersection(struct HitData* hit, struct Ray ray){
const float4 positions[] = {
(float4)(0, 1, -2, 1.0f),
(float4)(0, -100.0f, 0, 100.0f),
(float4)(0, 1, 0, 1.0f),
(float4)(0, 1, 5, 1.0f),
(float4)(10, 5, 5, 1.0f),
};
hit->length = 100000000.0f;
int hasHit = 0;
for (int i = 0; i < 5; i++){
if (checkIfHit( hit, ray, (float3)(positions[i].x, positions[i].y, positions[i].z), positions[i].w, 0.001f, hit->length )){
hasHit = i+1;
}
}
return hasHit;
}
float4 raycastI(struct Ray ray){
const float4 colors[] = {
(float4)(1.0f, 0.0f, 0.0f, 1.0f),
(float4)(0.0f,1.0f,0.0f, 1.0f),
(float4)(0.0f,0.0f,1.0f, 1.0f),
(float4)(0.0f,1.0f,1.0f, 1.0f),
(float4)(1.0f,0.0f,1.0f, 1.0f)
};
struct Ray localRay = ray;
float4 localColor = (float4)(1.0f);
int x = get_global_id(0);
int y = get_global_id(1);
for (int _ = 0; _ < MAX_DEPTH; _++){
struct HitData hit;
int hitIndex = checkWorldForIntersection(&hit, localRay);
if ( hitIndex ){
if (scatter(&localRay, hit, _)){
localColor = localColor * colors[hitIndex-1];
} else {
localColor = (float4)(0.0,0.0,0.0,0.0);
break;
}
} else {
// since we didn't hit, we hit the sky.
localColor = localColor * (float4)(0.5, 0.7, 1.0, 1.0);
// if we don't hit we cannot keep looping.
break;
}
localColor += localColor;
}
return localColor;
}
__kernel void raycast(__write_only image2d_t outputImage, __global unsigned char* objects, __global struct Vec* cameraData) {
unsigned long currentByte = 0;
int x = get_global_id(0);
int y = get_global_id(1);
float4 color = (float4)(0.0);
struct HitData hitData1;
if (checkIfHit(&hitData1, projectRay(x, y), (float4)(0.0f, -10, 0.0f, 0.0f), 10, 0, 1000)){
color = (float4)(hitData1.normal.x, hitData1.normal.y, hitData1.normal.z, 1.0f);
}
struct HitData hitData2;
if (checkIfHit(&hitData2, projectRay(x, y), (float4)(0.0f, 0, 0.0f, 0.0f), 1, 0, 1000)){
color = (float4)(0.0f, 1.0f, 0.0f, 1.0f);
for (int i = 0; i < MAX_PER_PIXEL; i++){
pcg6432_state state;
unsigned long seed = x * y * i;
pcg6432_seed(&state, seed);
color = color + raycastI(projectRay(cameraData, x + pcg6432_float(state), y + pcg6432_float(state)));
}
write_imagef(outputImage, (int2)(x, y), (float4)(sqrt(color.x), sqrt(color.y), sqrt(color.z), 1.0f));
float scaleFactor = 1.0 / MAX_PER_PIXEL;
write_imagef(outputImage, (int2)(x, y), (float4)(sqrt(color.x * scaleFactor), sqrt(color.y * scaleFactor), sqrt(color.z * scaleFactor), 1.0f));
//pcg6432_state state;
//unsigned long seed = x * y;
//pcg6432_seed(&state, seed);
//write_imagef(outputImage, (int2)(x, y), (float4)(randomVector(state), 1.0f));
}

16
Step 3/src/engine/main.cpp
View File

@ -207,9 +207,10 @@ int main(int argc, char** args) {
Raytracing::RayCaster rayCaster{camera, image, world, parser};
Texture mainImage(&image);
OpenClRaytracer openClRaytracer{parser.getOptionValue("--resources") + "opencl/raytracer.cl", image, camera, world};
openClRaytracer.run();
#ifdef COMPILE_OPENCL
OpenClRaytracer openClRaytracer{parser.getOptionValue("--resources") + "opencl/sphereray.cl", image, camera, world};
int frameCounter = 0;
#endif
Shader shader("../resources/shaders/basic.vs", "../resources/shaders/basic.fs");
Raytracing::DisplayRenderer renderer{*window, mainImage, world, shader, worldShader, rayCaster, parser, camera};
@ -217,6 +218,15 @@ int main(int argc, char** args) {
window->beginUpdate();
renderer.draw();
#ifdef COMPILE_OPENCL
openClRaytracer.updateCameraInformation();
//frameCounter++;
//if (frameCounter % 5 == 0) {
openClRaytracer.run();
// frameCounter = 0;
//}
#endif
glPolygonMode(GL_FRONT_AND_BACK, GL_LINE);
world.drawBVH(worldShader);

12
Step 3/src/engine/raytracing.cpp
View File

@ -10,8 +10,10 @@
#include <config.h>
#ifdef USE_MPI
#include <mpi/mpi.h>
#include <engine/mpi.h>
#else
#ifdef USE_OPENMP
#include <omp.h>
@ -41,10 +43,6 @@ namespace Raytracing {
imageOrigin = position - horizontalAxis / 2 - verticalAxis / 2 - w;
}
void Camera::setRotation(const PRECISION_TYPE yaw, const PRECISION_TYPE pitch) {
// TODO:
}
Mat4x4 Camera::view(PRECISION_TYPE yaw, PRECISION_TYPE pitch) {
Mat4x4 view;
@ -204,7 +202,7 @@ namespace Raytracing {
setupQueue(partitionScreen(threads));
#ifdef USE_OPENMP
ilog << "Running OpenMP\n";
#pragma omp parallel num_threads(threads+1) default(none) shared(threads)
#pragma omp parallel num_threads(threads+1) default(none) shared(threads)
{
int threadID = omp_get_thread_num();
// an attempt at making the omp command non-blocking.
@ -248,6 +246,7 @@ namespace Raytracing {
}
void RayCaster::runMPI(std::queue<RaycasterImageBounds> bounds) {
#ifdef USE_MPI
ilog << "Running MPI\n";
dlog << "We have " << bounds.size() << " bounds currently pending!\n";
while (!bounds.empty()) {
@ -259,8 +258,9 @@ namespace Raytracing {
}
bounds.pop();
}
#ifdef USE_MPI
dlog << "Finished running MPI on " << currentProcessID << "\n";
#else
flog << "Not compiled with MPI!\n";
#endif
}

16
Step 3/src/engine/util/loaders.cpp
View File

@ -52,12 +52,15 @@ namespace Raytracing {
throw std::runtime_error("");
}
try {
// filter out the > | " at the end of the include statement.
std::string file = statement1.empty() ? statement2[1].substr(0, statement2[1].size() - 1) : statement1[1].substr(
0,
statement1[1].size() -
1
);
// filter out the > or " at the end of the include statement.
std::string file;
if (statement1.size() <= 1) {
// since " is one character, we don't need to substring since split already handles it.
file = statement2[1];
} else {
tlog << statement1[1];
file = statement1[1].substr(0, statement1[1].size() - 1);
}
tlog << "Recusing into " << (pathOnly + "/" + file) << "\n";
@ -105,6 +108,7 @@ namespace Raytracing {
stringStream << "\n";
}
//tlog << stringStream.str();
return stringStream.str();
}

102
Step 3/src/graphics/graphics.cpp
View File

@ -61,8 +61,9 @@ namespace Raytracing {
return imager;
}
#ifdef USE_GLFW
XWindow::XWindow(int width, int height): m_displayWidth(width), m_displayHeight(height) {
#ifdef USE_GLFW
XWindow::XWindow(int width, int height):
m_displayWidth(width), m_displayHeight(height) {
// OpenGL 4.6 is like 5 years old at this point and most systems support it
glfwWindowHint(GLFW_CONTEXT_VERSION_MAJOR, 4);
glfwWindowHint(GLFW_CONTEXT_VERSION_MINOR, 6);
@ -121,7 +122,8 @@ namespace Raytracing {
IMGUI_CHECKVERSION();
ImGui::CreateContext();
ImGuiIO& io = ImGui::GetIO(); (void)io;
ImGuiIO& io = ImGui::GetIO();
(void) io;
ImGui::StyleColorsDark();
// create ImGUI GLFW instance and install the callbacks
@ -129,7 +131,7 @@ namespace Raytracing {
ImGui_ImplOpenGL3_Init("#version 130");
int version = gladLoadGL(glfwGetProcAddress);
if(!version)
if (!version)
throw std::runtime_error("Unable to load Glad GL!\n");
glfwShowWindow(window);
@ -179,9 +181,9 @@ namespace Raytracing {
lastFrameTime = currentFrameTime;
frameTimeMs = delta / 1000000.0;
frameTimeS = delta / 1000000000.0;
fps = 1000/frameTimeMs;
fps = 1000 / frameTimeMs;
}
#else
#else
XWindow::XWindow(int width, int height): m_width(width), m_height(height) {
// open the DEFAULT display. We don't want to open a specific screen as that is annoying.
dlog << "Creating X11 display!\n";
@ -463,7 +465,7 @@ namespace Raytracing {
XDestroyWindow(display, window);
XCloseDisplay(display);
}
#endif
#endif
XWindow::~XWindow() {
closeWindow();
deleteKeys();
@ -480,40 +482,7 @@ namespace Raytracing {
static int raysPerPixel = 50;
static float yaw = 0, pitch = 0;
void DisplayRenderer::draw() {
if (RTSignal->haltExecution){m_window.closeWindow();}
if (Input::isKeyDown(GLFW_KEY_ESCAPE) && Input::isState(GLFW_KEY_ESCAPE))
m_window.setMouseGrabbed(!m_window.isMouseGrabbed());
DebugUI::render([this]() -> void {
if (ImGui::Button("Start") && !started){
started = true;
RTSignal->haltRaytracing = false;
ilog << "Running raycaster!\n";
// we don't actually have to check for --single since it's implied to be default true.
int threads = std::stoi(m_parser.getOptionValue("--threads"));
if (m_parser.hasOption("--mpi")) {
//m_raycaster.runMPI(raycaster.partitionScreen());
} else if(m_parser.hasOption("--openmp")){
m_raycaster.runOpenMP(threads);
} else {
m_raycaster.runSTDThread(threads);
}
}
if (ImGui::Checkbox("Pause", &RTSignal->pauseRaytracing)){}
if (ImGui::Button("Stop") && started){
RTSignal->haltRaytracing = true;
started = false;
m_raycaster.deleteThreads();
}
ImGui::NewLine();
ImGui::InputInt("Max Ray Bounce", &maxRayBounce);
ImGui::InputInt("Rays Per Pixel", &raysPerPixel);
m_raycaster.updateRayInfo(maxRayBounce, raysPerPixel);
ImGui::Checkbox("Debug", &debug);
});
if (debug){
std::pair<Mat4x4, Mat4x4> DisplayRenderer::getCameraMatrices() {
auto projection = m_camera.project();
if (m_window.isMouseGrabbed()) {
yaw += (float) Input::getMouseDelta().x * (1000.0f / ImGui::GetIO().Framerate / 1000.0f) * 3;
@ -561,6 +530,49 @@ namespace Raytracing {
m_camera.setPosition(m_camera.getPosition() + Vec4{deltaX, deltaY, deltaZ});
}
auto view = m_camera.view(yaw, pitch);
return {projection, view};
}
void DisplayRenderer::draw() {
if (RTSignal->haltExecution) { m_window.closeWindow(); }
if (Input::isKeyDown(GLFW_KEY_ESCAPE) && Input::isState(GLFW_KEY_ESCAPE))
m_window.setMouseGrabbed(!m_window.isMouseGrabbed());
DebugUI::render([this]() -> void {
if (ImGui::Button("Start") && !started) {
started = true;
RTSignal->haltRaytracing = false;
ilog << "Running raycaster!\n";
// we don't actually have to check for --single since it's implied to be default true.
int threads = std::stoi(m_parser.getOptionValue("--threads"));
if (m_parser.hasOption("--mpi")) {
//m_raycaster.runMPI(raycaster.partitionScreen());
} else if (m_parser.hasOption("--openmp")) {
m_raycaster.runOpenMP(threads);
} else {
m_raycaster.runSTDThread(threads);
}
}
if (ImGui::Checkbox("Pause", &RTSignal->pauseRaytracing)) {}
if (ImGui::Button("Stop") && started) {
RTSignal->haltRaytracing = true;
started = false;
m_raycaster.deleteThreads();
}
ImGui::NewLine();
ImGui::InputInt("Max Ray Bounce", &maxRayBounce);
ImGui::InputInt("Rays Per Pixel", &raysPerPixel);
m_raycaster.updateRayInfo(maxRayBounce, raysPerPixel);
ImGui::Checkbox("Debug", &debug);
});
// we want to be able to move around, and the camera matrix functions automatically recalculate image region & projection data.
if (m_parser.hasOption("--gpu")) {
getCameraMatrices();
}
if (debug) {
// if (Input::isKeyDown(GLFW_KEY_E) && Input::isState(GLFW_KEY_E)) {
// auto ray = m_camera.projectRay((PRECISION_TYPE) m_window.displayWidth() / 2, (PRECISION_TYPE) m_window.displayHeight() / 2);
//
@ -577,12 +589,12 @@ namespace Raytracing {
// }
// if (Input::isKeyDown(GLFW_KEY_R) && Input::isState(GLFW_KEY_R))
// m_world.getBVH()->resetNodes();
auto view = m_camera.view(yaw, pitch);
m_worldShader.setMatrix("projectMatrix", projection);
m_worldShader.setMatrix("viewMatrix", view);
auto matrices = getCameraMatrices();
m_worldShader.setMatrix("projectMatrix", matrices.first);
m_worldShader.setMatrix("viewMatrix", matrices.second);
m_worldShader.use();
auto objs = m_world.getObjectsInWorld();
for (auto obj : objs) {
for (auto obj: objs) {
if (obj->getVAO() != nullptr) {
obj->getVAO()->bind();
obj->getVAO()->draw(m_worldShader, {obj->getPosition()});

73
Step 3/src/opencl/open_ray_tracing.cpp
View File

@ -3,6 +3,8 @@
* Copyright (c) 2022 Brett Terpstra. All Rights Reserved.
*/
#include <opencl/open_ray_tracing.h>
#include <cstddef>
#include "engine/util/loaders.h"
namespace Raytracing {
@ -13,7 +15,7 @@ namespace Raytracing {
constexpr size_t vecBytes = sizeof(float) * 4;
OpenClRaytracer::OpenClRaytracer(const std::string& programLocation, Image& image, Camera& camera, World& world):
image(image) {
image(image), camera(camera) {
auto objectsInWorld = world.getObjectsInWorld();
objectCount = 0;
// pre-calculate the space needed for objects, since every object must statically store these number of triangles, even if not used.
@ -31,7 +33,7 @@ namespace Raytracing {
// 3 vectors per object, 1 size type per object, maxTriangleSize triangles.
size_t totalWorldBytes =
(vecBytes * 3) * objectCount + sizeof(unsigned long) * objectCount + triangleNumOfBytes * maxTriangleSize * objectCount;
auto objectBuffer = createObjectBuffer(objectsInWorld, totalWorldBytes);
//auto objectBuffer = createObjectBuffer(objectsInWorld, totalWorldBytes);
ShaderLoader::define("maxTriangleCount", std::to_string(maxTriangleSize));
ShaderLoader::define("objectCount", std::to_string(objectCount));
@ -40,30 +42,26 @@ namespace Raytracing {
// load up information about the camera. Since these don't generally chance at runtime we can load them up at compile time
// however this means that changes made in debug mode do not transfer.
auto origin = camera.getImageOrigin();
ShaderLoader::define(
"imageOrigin",
"(float4)(" + std::to_string(origin.x()) + ", " + std::to_string(origin.y()) + ", " + std::to_string(origin.z()) + ", " +
std::to_string(origin.w()) + ")"
);
auto horiz = camera.getHorizontalAxis();
ShaderLoader::define(
"horizontalAxis",
"(float4)(" + std::to_string(horiz.x()) + ", " + std::to_string(horiz.y()) + ", " + std::to_string(horiz.z()) + ", " +
std::to_string(horiz.w()) + ")"
);
auto vert = camera.getHorizontalAxis();
ShaderLoader::define(
"verticalAxis",
"(float4)(" + std::to_string(vert.x()) + ", " + std::to_string(vert.y()) + ", " + std::to_string(vert.z()) + ", " +
std::to_string(vert.w()) + ")"
);
auto pos = camera.getPosition();
ShaderLoader::define(
"cameraPosition",
"(float4)(" + std::to_string(pos.x()) + ", " + std::to_string(pos.y()) + ", " + std::to_string(pos.z()) + ", " +
std::to_string(pos.w()) + ")"
);
// auto origin = camera.getImageOrigin();
// ShaderLoader::define(
// "imageOrigin",
// "(float3)(" + std::to_string(origin.x()) + ", " + std::to_string(origin.y()) + ", " + std::to_string(origin.z()) + ")"
// );
// auto horiz = camera.getHorizontalAxis();
// ShaderLoader::define(
// "horizontalAxis",
// "(float3)(" + std::to_string(horiz.x()) + ", " + std::to_string(horiz.y()) + ", " + std::to_string(horiz.z()) + ")"
// );
// auto vert = camera.getVerticalAxis();
// ShaderLoader::define(
// "verticalAxis",
// "(float3)(" + std::to_string(vert.x()) + ", " + std::to_string(vert.y()) + ", " + std::to_string(vert.z()) + ")"
// );
// auto pos = camera.getPosition();
// ShaderLoader::define(
// "cameraPosition",
// "(float3)(" + std::to_string(pos.x()) + ", " + std::to_string(pos.y()) + ", " + std::to_string(pos.z()) + ")"
// );
program = new CLProgram(programLocation);
OpenCL::createCLProgram(*program);
@ -72,19 +70,13 @@ namespace Raytracing {
program->createImage("outputImage", image.getWidth(), image.getHeight());
program->createBuffer("objects", CL_MEM_READ_WRITE, totalWorldBytes);
program->createBuffer("randoms", CL_MEM_READ_ONLY, image.getWidth() * vecBytes);
// the raytracing algorithm needs a good supply of random numbers. This creates it for us on the CPU as to prevent needing to generate on GPU.
auto randomsBufferBytes = new unsigned char[image.getWidth() * vecBytes];
size_t currentPos = 0;
for (int i = 0; i < image.getWidth(); i++) {
MemoryConvert::writeVectorBytes(randomsBufferBytes, currentPos, Raytracing::RayCaster::randomUnitVector().normalize());
}
program->writeBuffer("randoms", image.getWidth() * vecBytes, randomsBufferBytes);
storeObjects(objectBuffer, totalWorldBytes);
program->createBuffer("cameraData", CL_MEM_READ_WRITE, sizeof(float) * 3 * 4);
//storeObjects(objectBuffer, totalWorldBytes);
program->setKernelArgument("raycast", "outputImage", 0);
program->setKernelArgument("raycast", "objects", 1);
program->setKernelArgument("raycast", "randoms", 2);
program->setKernelArgument("raycast", "cameraData", 2);
updateCameraInformation();
}
OpenClRaytracer::~OpenClRaytracer() {
@ -167,4 +159,13 @@ namespace Raytracing {
// }
return buffer;
}
void OpenClRaytracer::updateCameraInformation() {
unsigned char buffer[sizeof(float) * 3 * 4];
size_t currentIndex = 0;
MemoryConvert::writeVectorBytes(buffer, currentIndex, camera.getPosition());
MemoryConvert::writeVectorBytes(buffer, currentIndex, camera.getVerticalAxis());
MemoryConvert::writeVectorBytes(buffer, currentIndex, camera.getHorizontalAxis());
MemoryConvert::writeVectorBytes(buffer, currentIndex, camera.getImageOrigin());
program->writeBuffer("cameraData", sizeof(float) * 3 * 4, buffer);
}
}