diff --git a/include/CL_Wrapper.h b/include/CL_Wrapper.h index 99bd728..8348536 100644 --- a/include/CL_Wrapper.h +++ b/include/CL_Wrapper.h @@ -42,17 +42,20 @@ public: int run_kernel(std::string kernel_name); + bool assert(int error_code, std::string function_name); + cl_device_id getDeviceID(); cl_platform_id getPlatformID(); cl_context getContext(); cl_kernel getKernel(std::string kernel_name); + cl_command_queue getCommandQueue(); private: int error = 0; bool initialized = false; - bool assert(int error_code, std::string function_name); + cl_platform_id platform_id; cl_device_id device_id; cl_context context; diff --git a/kernels/minimal_kernel.c b/kernels/minimal_kernel.c index 359162c..db32d64 100644 --- a/kernels/minimal_kernel.c +++ b/kernels/minimal_kernel.c @@ -3,13 +3,17 @@ __kernel void min_kern( global char* map, global int3* map_dim, global int2* resolution, - global float3* projection_matrix + global float3* projection_matrix, + global float3* cam_dir, + global float3* cam_pos ){ size_t id = get_global_id(0); //printf("%i %c -- ", id, map[id]); //printf("%i, %i, %i\n", map_dim->x, map_dim->y, map_dim->z); - printf("\n%i\nX: %f\nY: %f\nZ: %f\n", id, projection_matrix[id].x, projection_matrix[id].y, projection_matrix[id].z); + //printf("\n%i\nX: %f\nY: %f\nZ: %f\n", id, projection_matrix[id].x, projection_matrix[id].y, projection_matrix[id].z); + //printf("%f, %f, %f\n", cam_dir->x, cam_dir->y, cam_dir->z); + //printf("%f, %f, %f\n", cam_pos->x, cam_pos->y, cam_pos->z); } \ No newline at end of file diff --git a/src/CL_Wrapper.cpp b/src/CL_Wrapper.cpp index 9fb2346..2427d85 100644 --- a/src/CL_Wrapper.cpp +++ b/src/CL_Wrapper.cpp @@ -254,7 +254,8 @@ int CL_Wrapper::run_kernel(std::string kernel_name){ cl_device_id CL_Wrapper::getDeviceID(){ return device_id; }; cl_platform_id CL_Wrapper::getPlatformID(){ return platform_id; }; cl_context CL_Wrapper::getContext(){ return context; }; -cl_kernel CL_Wrapper::getKernel(std::string kernel_name ){ return kernel_map.at(kernel_name); } +cl_kernel CL_Wrapper::getKernel(std::string kernel_name ){ return kernel_map.at(kernel_name); }; +cl_command_queue CL_Wrapper::getCommandQueue(){ return command_queue; }; bool CL_Wrapper::assert(int error_code, std::string function_name){ diff --git a/src/main.cpp b/src/main.cpp index c87ca77..0e7a87f 100644 --- a/src/main.cpp +++ b/src/main.cpp @@ -15,6 +15,7 @@ #include #elif defined TARGET_OS_MAC +#include # include # include #include @@ -33,142 +34,195 @@ const int WINDOW_Y = 150; -int main(){ - CL_Wrapper c; - c.acquire_platform_and_device(); - c.create_shared_context(); - c.create_command_queue(); - c.compile_kernel("../kernels/kernel.c", true, "hello"); - c.compile_kernel("../kernels/minimal_kernel.c", true, "min_kern"); +float elap_time(){ + static std::chrono::time_point start; + static bool started = false; - std::string in = "hello!!!!!!!!!!!!!!!!!!!!!"; - cl_mem buff = clCreateBuffer( - c.getContext(), CL_MEM_READ_WRITE | CL_MEM_COPY_HOST_PTR, - sizeof(char) * 128, &in[0], NULL - ); + if (!started){ + start = std::chrono::system_clock::now(); + started = true; + } - char map[100 * 100 * 100]; + std::chrono::time_point now = std::chrono::system_clock::now(); + std::chrono::duration elapsed_time = now - start; + return elapsed_time.count(); +} - for (int i = 0; i < 100*100*100; i++){ - map[i] = '+'; - } +sf::Sprite window_sprite; +sf::Texture window_texture; - map[0] = 'a'; +// Y: -1.57 is straight up +// Y: 1.57 is straight down - cl_mem map_buff = clCreateBuffer( - c.getContext(), CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, - sizeof(char) * 100*100*100, map, NULL - ); - int dim[3] = {101, 100, 99}; +int main() { - cl_mem dim_buff = clCreateBuffer( - c.getContext(), CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, - sizeof(int) * 3, dim, NULL - ); + sf::RenderWindow window(sf::VideoMode(WINDOW_X, WINDOW_Y), "SFML"); - int res[2] = {100, 99}; - cl_mem res_buff = clCreateBuffer( - c.getContext(), CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, - sizeof(int) * 2, res, NULL - ); + sf::Sprite s; + sf::Texture t; - double y_increment_radians = DegreesToRadians(50.0 / res[1]); - double x_increment_radians = DegreesToRadians(80.0 / res[0]); + { + CL_Wrapper c; + c.acquire_platform_and_device(); + c.create_shared_context(); + c.create_command_queue(); - // SFML 2.4 has Vector4 datatypes....... - sf::Vector3f* view_plane_vectors = new sf::Vector3f[res[0] * res[1]]; - for (int y = -res[1] / 2 ; y < res[1] / 2; y++) { - for (int x = -res[0] / 2; x < res[0] / 2; x++) { + c.compile_kernel("../kernels/kernel.c", true, "hello"); + c.compile_kernel("../kernels/minimal_kernel.c", true, "min_kern"); - // The base ray direction to slew from - sf::Vector3f ray(1, 0, 0); + std::string in = "hello!!!!!!!!!!!!!!!!!!!!!"; + cl_mem buff = clCreateBuffer( + c.getContext(), CL_MEM_READ_WRITE | CL_MEM_COPY_HOST_PTR, + sizeof(char) * 128, &in[0], NULL + ); - // Y axis, pitch - ray = sf::Vector3f( - ray.z * sin(y_increment_radians * y) + ray.x * cos(y_increment_radians * y), - ray.y, - ray.z * cos(y_increment_radians * y) - ray.x * sin(y_increment_radians * y) - ); + char map[100 * 100 * 100]; - // Z axis, yaw - ray = sf::Vector3f( - ray.x * cos(x_increment_radians * x) - ray.y * sin(x_increment_radians * x), - ray.x * sin(x_increment_radians * x) + ray.y * cos(x_increment_radians * x), - ray.z - ); + for (int i = 0; i < 100 * 100 * 100; i++) { + map[i] = '+'; + } - int index = (x + res[0] / 2) + res[0] * (y + res[1] / 2); - view_plane_vectors[index] = Normalize(ray); + map[0] = 'a'; + + cl_mem map_buff = clCreateBuffer( + c.getContext(), CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, + sizeof(char) * 100 * 100 * 100, map, NULL + ); + + int dim[3] = {101, 100, 99}; + + cl_mem dim_buff = clCreateBuffer( + c.getContext(), CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, + sizeof(int) * 3, dim, NULL + ); + + int res[2] = {100, 99}; + + cl_mem res_buff = clCreateBuffer( + c.getContext(), CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, + sizeof(int) * 2, res, NULL + ); + + double y_increment_radians = DegreesToRadians(50.0 / res[1]); + double x_increment_radians = DegreesToRadians(80.0 / res[0]); + + // SFML 2.4 has Vector4 datatypes....... + + float view_matrix[res[0] * res[1] * 4]; + for (int y = -res[1] / 2; y < res[1] / 2; y++) { + for (int x = -res[0] / 2; x < res[0] / 2; x++) { + + // The base ray direction to slew from + sf::Vector3f ray(1, 0, 0); + + // Y axis, pitch + ray = sf::Vector3f( + ray.z * sin(y_increment_radians * y) + ray.x * cos(y_increment_radians * y), + ray.y, + ray.z * cos(y_increment_radians * y) - ray.x * sin(y_increment_radians * y) + ); + + // Z axis, yaw + ray = sf::Vector3f( + ray.x * cos(x_increment_radians * x) - ray.y * sin(x_increment_radians * x), + ray.x * sin(x_increment_radians * x) + ray.y * cos(x_increment_radians * x), + ray.z + ); + + int index = (x + res[0] / 2) + res[0] * (y + res[1] / 2); + ray = Normalize(ray); + view_matrix[index * 4 + 0] = ray.x; + view_matrix[index * 4 + 1] = ray.y; + view_matrix[index * 4 + 2] = ray.z; + view_matrix[index * 4 + 3] = 0; + } } - } - int ind = 1; - std::cout << "\nX: " << view_plane_vectors[ind].x - << "\nY: " << view_plane_vectors[ind].y - << "\nZ: " << view_plane_vectors[ind].z; +// int ind = 4; +// std::cout << "\nX: " << view_matrix[ind] +// << "\nY: " << view_matrix[ind + 1] +// << "\nZ: " << view_matrix[ind + 2] +// << "\npad: " << view_matrix[ind + 3]; +// +// std::cout << "\n======================" << std::endl; - std::cout << "\n======================" << std::endl; + cl_mem view_matrix_buff = clCreateBuffer( + c.getContext(), CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, + sizeof(float) * 3 * res[0] * res[1], view_matrix, NULL + ); - cl_mem view_matrix_buff = clCreateBuffer( - c.getContext(), CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, - sizeof(float) * 3 * res[0] * res[1], &view_plane_vectors[0], NULL - ); - c.store_buffer(buff, "buffer_1"); - c.store_buffer(map_buff, "map_buffer"); - c.store_buffer(dim_buff, "dim_buffer"); - c.store_buffer(res_buff, "res_buffer"); - c.store_buffer(view_matrix_buff, "view_matrix_buffer"); + float cam_dir[4] = {1, 0, 0, 0}; - c.set_kernel_arg("min_kern", 0, "buffer_1"); - c.set_kernel_arg("min_kern", 1, "map_buffer"); - c.set_kernel_arg("min_kern", 2, "dim_buffer"); - c.set_kernel_arg("min_kern", 3, "res_buffer"); - c.set_kernel_arg("min_kern", 4, "view_matrix_buffer"); + cl_mem cam_dir_buff = clCreateBuffer( + c.getContext(), CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, + sizeof(float) * 4, cam_dir, NULL + ); - c.run_kernel("min_kern"); + float cam_pos[4] = {25, 25, 25, 0}; + cl_mem cam_pos_buff = clCreateBuffer( + c.getContext(), CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, + sizeof(float) * 4, cam_pos, NULL + ); -}; + c.store_buffer(buff, "buffer_1"); + c.store_buffer(map_buff, "map_buffer"); + c.store_buffer(dim_buff, "dim_buffer"); + c.store_buffer(res_buff, "res_buffer"); + c.store_buffer(view_matrix_buff, "view_matrix_buffer"); + c.store_buffer(cam_dir_buff, "cam_dir_buffer"); + c.store_buffer(cam_pos_buff, "cam_pos_buffer"); + c.set_kernel_arg("min_kern", 0, "buffer_1"); + c.set_kernel_arg("min_kern", 1, "map_buffer"); + c.set_kernel_arg("min_kern", 2, "dim_buffer"); + c.set_kernel_arg("min_kern", 3, "res_buffer"); + c.set_kernel_arg("min_kern", 4, "view_matrix_buffer"); + c.set_kernel_arg("min_kern", 5, "cam_dir_buffer"); + c.set_kernel_arg("min_kern", 6, "cam_pos_buffer"); + c.run_kernel("min_kern"); + unsigned char* pixel_array = new sf::Uint8[WINDOW_X * WINDOW_Y * 4]; -float elap_time(){ - static std::chrono::time_point start; - static bool started = false; + for (int i = 0; i < 100 * 100 * 4; i += 4) { - if (!started){ - start = std::chrono::system_clock::now(); - started = true; - } + pixel_array[i] = i % 255; // R? + pixel_array[i + 1] = 70; // G? + pixel_array[i + 2] = 100; // B? + pixel_array[i + 3] = 100; // A? + } - std::chrono::time_point now = std::chrono::system_clock::now(); - std::chrono::duration elapsed_time = now - start; - return elapsed_time.count(); -} + t.create(100, 100); + t.update(pixel_array); -sf::Sprite window_sprite; -sf::Texture window_texture; + int error; -// Y: -1.57 is straight up -// Y: 1.57 is straight down + cl_mem image_buff = clCreateFromGLTexture(c.getContext(), CL_MEM_WRITE_ONLY, GL_TEXTURE_2D, 0, t.getNativeHandle(), &error); + if (c.assert(error, "clCreateFromGLTexture")) + return -1; + error = clEnqueueAcquireGLObjects(c.getCommandQueue(), 1, &image_buff, 0, 0, 0); + if (c.assert(error, "clEnqueueAcquireGLObjects")) + return -1; -int main0() { + //c.run_kernel("min_kern"); - // Initialize the render window - Curses curse(sf::Vector2i(5, 5), sf::Vector2i(WINDOW_X, WINDOW_Y)); - sf::RenderWindow window(sf::VideoMode(WINDOW_X, WINDOW_Y), "SFML"); - + error = clEnqueueReleaseGLObjects(c.getCommandQueue(), 1, &image_buff, 0, NULL, NULL); + if (c.assert(error, "clEnqueueReleaseGLObjects")) + return -1; + + s.setTexture(t); + } // The step size in milliseconds between calls to Update() // Lets set it to 16.6 milliseonds (60FPS) float step_size = 0.0166f; @@ -183,7 +237,7 @@ int main0() { fps_counter fps; // ============================= RAYCASTER SETUP ================================== - + // Setup the sprite and texture window_texture.create(WINDOW_X, WINDOW_Y); window_sprite.setPosition(0, 0); @@ -264,21 +318,6 @@ int main0() { cam_pos.y += cam_vec.y / 1.0; cam_pos.z += cam_vec.z / 1.0; -// if (cam_vec.x > 0.0f) -// cam_vec.x -= 0.1; -// else if (cam_vec.x < 0.0f) -// cam_vec.x += 0.1; -// -// if (cam_vec.y > 0.0f) -// cam_vec.y -= 0.1; -// else if (cam_vec.y < 0.0f) -// cam_vec.y += 0.1; -// -// if (cam_vec.z > 0.0f) -// cam_vec.z -= 0.1; -// else if (cam_vec.z < 0.0f) -// cam_vec.z += 0.1; - std::cout << cam_vec.x << " : " << cam_vec.y << " : " << cam_vec.z << std::endl; @@ -292,31 +331,17 @@ int main0() { while ((accumulator_time - step_size) >= step_size) { accumulator_time -= step_size; - // Update cycle - curse.Update(delta_time); - - } // Fps cycle // map->moveLight(sf::Vector2f(0.3, 0)); - + window.clear(sf::Color::Black); // Cast the rays and get the image sf::Color* pixel_colors = ray_caster.CastRays(cam_dir, cam_pos); - for (int i = 0; i < WINDOW_X * WINDOW_Y; i++) { - - Curses::Tile t(sf::Vector2i(i % WINDOW_X, i / WINDOW_X)); - Curses::Slot s(L'\u0045', pixel_colors[i], sf::Color::Black); - t.push_back(s); - curse.setTile(t); - - } - - // Cast it to an array of Uint8's auto out = (sf::Uint8*)pixel_colors; @@ -324,18 +349,15 @@ int main0() { window_sprite.setTexture(window_texture); window.draw(window_sprite); - - curse.Render(); - // Give the frame counter the frame time and draw the average frame time fps.frame(delta_time); fps.draw(&window); - window.display(); - + window.draw(s); + + window.display(); } return 0; - }