From a7234675cb19b542d1e94d01842dbc41095146a0 Mon Sep 17 00:00:00 2001 From: mitchellhansen Date: Wed, 31 Aug 2016 21:47:56 -0700 Subject: [PATCH] Continuing on adding argument handling. Added a small sample texture and sprite which can be handed over to either gl or cl and then handed back. It can then be rendered. Changed to just an array of floats for the view matrix --- include/CL_Wrapper.h | 5 +- kernels/minimal_kernel.c | 8 +- src/CL_Wrapper.cpp | 3 +- src/main.cpp | 314 +++++++++++++++++++++------------------ 4 files changed, 180 insertions(+), 150 deletions(-) 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,111 +34,6 @@ 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"); - - 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 - ); - - char map[100 * 100 * 100]; - - for (int i = 0; i < 100*100*100; i++){ - map[i] = '+'; - } - - 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....... - 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++) { - - // 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); - view_plane_vectors[index] = Normalize(ray); - } - } - - int ind = 1; - std::cout << "\nX: " << view_plane_vectors[ind].x - << "\nY: " << view_plane_vectors[ind].y - << "\nZ: " << view_plane_vectors[ind].z; - - 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_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"); - - 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.run_kernel("min_kern"); - - -}; - - - - float elap_time(){ @@ -161,14 +57,172 @@ sf::Texture window_texture; // Y: 1.57 is straight down -int main0() { +int main() { - // 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"); - + sf::Sprite s; + sf::Texture t; + + { + 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"); + + 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 + ); + + char map[100 * 100 * 100]; + + for (int i = 0; i < 100 * 100 * 100; i++) { + map[i] = '+'; + } + + 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 = 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; + + 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 + ); + + + float cam_dir[4] = {1, 0, 0, 0}; + + cl_mem cam_dir_buff = clCreateBuffer( + c.getContext(), CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, + sizeof(float) * 4, cam_dir, NULL + ); + + 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]; + + for (int i = 0; i < 100 * 100 * 4; i += 4) { + + pixel_array[i] = i % 255; // R? + pixel_array[i + 1] = 70; // G? + pixel_array[i + 2] = 100; // B? + pixel_array[i + 3] = 100; // A? + } + + t.create(100, 100); + t.update(pixel_array); + + int error; + + 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; + + //c.run_kernel("min_kern"); + + 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.draw(s); + window.display(); - - } return 0; - }