From d1bd4ce6675049a2e56b66edbc4a0cb11e0478b4 Mon Sep 17 00:00:00 2001 From: MitchellHansen Date: Fri, 4 Nov 2016 03:28:23 -0700 Subject: [PATCH] The new rendering method now works on the full compat case. Added a kernel to help test opencl data passing renamed the kernels, buffers, etc. --- CMakeLists.txt | 7 +- include/Curses.h | 138 ----------------------- include/Hardware_Caster.h | 12 +- include/RayCaster.h | 14 ++- include/util.hpp | 6 + kernels/kernel.cl | 125 --------------------- kernels/print_arguments.cl | 32 ++++++ kernels/ray_caster_kernel.cl | 2 +- src/CL_Wrapper.cpp | 3 +- src/Curses.cpp | 187 -------------------------------- src/Hardware_Caster.cpp | 87 ++++++++++++--- src/RayCaster.cpp | 205 ++++++++++++++++++----------------- src/main.cpp | 136 +++++------------------ 13 files changed, 268 insertions(+), 686 deletions(-) delete mode 100644 include/Curses.h delete mode 100644 kernels/kernel.cl create mode 100644 kernels/print_arguments.cl delete mode 100644 src/Curses.cpp diff --git a/CMakeLists.txt b/CMakeLists.txt index 0e7212e..af361dc 100644 --- a/CMakeLists.txt +++ b/CMakeLists.txt @@ -27,9 +27,12 @@ include_directories(${OpenCL_INCLUDE_DIRS}) include_directories(${OpenGL_INCLUDE_DIRS}) include_directories(include) -# Set the .cpp sources +# Set the sources, allows VS to filter them properly file(GLOB SOURCES "src/*.cpp") -add_executable(${PNAME} ${SOURCES} include/Renderer.cpp include/Renderer.h) +file(GLOB HEADERS "include/*.h" "include/*.hpp") +file(GLOB KERNELS "kernels/*.cl") + +add_executable(${PNAME} ${SOURCES} ${HEADERS} ${KERNELS}) # Link CL, GL, and SFML target_link_libraries (${PNAME} ${SFML_LIBRARIES} ${SFML_DEPENDENCIES}) diff --git a/include/Curses.h b/include/Curses.h deleted file mode 100644 index 0634101..0000000 --- a/include/Curses.h +++ /dev/null @@ -1,138 +0,0 @@ -#pragma once -#include -#include - -class Curses { -public: - - struct Slot { - Slot(wchar_t unicode_value_, - sf::Color font_color_, - sf::Color backfill_color_) : - unicode_value(unicode_value_), - font_color(font_color_), - backfill_color(backfill_color_) - {}; - - wchar_t unicode_value; - sf::Color font_color; - sf::Color backfill_color; - }; - - struct Tile { - public: - Tile(sf::Vector2i position_) : - blank_standby(L'\u0020', sf::Color::Transparent, sf::Color::Black), - position(position_) - { }; - - private: - Slot blank_standby; - - int index = 0; // What index in the vector are we. Backbone for blinking and scrolling - int ratio_counter = 0; // Secondary counter to hold index positions for (ratio) length of time - int ratio_value = 0; - - std::vector slot_stack; // The icon that aligns with the index - - sf::Vector2i position; // Position of the text, and backfill - - public: - void set_ratio(int ratio_) { - ratio_value = ratio_; - } - - sf::Vector2i getPosition() const { - return position; - } - - void push_back(Slot s) { - slot_stack.push_back(s); - } - - void clear_and_set(Slot s) { - slot_stack.clear(); - slot_stack.push_back(s); - } - - void clear() { - slot_stack.clear(); - } - - sf::Color current_font_color() { - if (slot_stack.size() > 0) - return slot_stack.at(index).font_color; - else - return blank_standby.font_color; - } - - sf::Color current_backfill_color() { - if (slot_stack.size() > 0) - return slot_stack.at(index).backfill_color; - else - return blank_standby.backfill_color; - } - - wchar_t current_unicode_value() { - if (slot_stack.size() > 0) - return slot_stack.at(index).unicode_value; - else - return blank_standby.unicode_value; - } - - void inc_index() { - if (index >= slot_stack.size() - 1) { - index = 0; - } - else if (ratio_counter == ratio_value) { - ratio_counter = 0; - index++; - } - else - ratio_counter++; - } - }; - - Curses(sf::Vector2i tile_size_, sf::Vector2i grid_dimensions); - ~Curses(); - - - void Update(double delta_time_); - void Render(); - - - void setTile(Tile tile_); - void setTiles(std::vector tiles_); // Can be seperate, non-adjacent tiles - - void Clear(); - - Tile* getTile(sf::Vector2i position_); - std::vector getTiles(sf::Vector2i start_, sf::Vector2i end_); - - void ResizeTiles(sf::Vector2i size_); - void ResizeTileGrid(sf::Vector2i grid_dimensions_); - - void setBlink(int ratio_, sf::Vector2i position_); - void setScroll(int ratio_, sf::Vector2i start_, sf::Vector2i end_); - void setScroll(int ratio_, sf::Vector2i start_, std::list tiles_); - - -private: - - sf::Vector2i grid_dimensions; - sf::Vector2i tile_pixel_dimensions; - - sf::RenderWindow window; - - std::vector tiles; - - sf::Font font; - - int multi_to_linear(sf::Vector2i position_) const; - sf::Vector2i linear_to_multi(int position_) const; - - void set_tile_ratio(int ratio_, sf::Vector2i tile_position_); - void append_slots(sf::Vector2i start_, std::list values_); - -}; - diff --git a/include/Hardware_Caster.h b/include/Hardware_Caster.h index 27bfed5..95cdf9c 100644 --- a/include/Hardware_Caster.h +++ b/include/Hardware_Caster.h @@ -6,6 +6,7 @@ #include #include + #ifdef linux #include #include @@ -15,8 +16,10 @@ #include #include #include -#include + +// Note: windows.h must be included before Gl/GL.h #include +#include #elif defined TARGET_OS_MAC # include @@ -47,12 +50,14 @@ public: // Both will create the view matrix, view res buffer void create_viewport(int width, int height, float v_fov, float h_fov) override; - void assign_light(std::string light_id, Light light) override; + void assign_lights(std::vector lights) override; void assign_map(Old_Map *map) override; void assign_camera(Camera *camera) override; + void validate() override; // draw will abstract the gl sharing and software rendering // methods of retrieving the screen buffer + void compute() override; void draw(sf::RenderWindow* window) override; private: @@ -78,6 +83,8 @@ private: int run_kernel(std::string kernel_name, const int work_size); + void print_kernel_arguments(); + bool assert(int error_code, std::string function_name); cl_device_id getDeviceID(); @@ -93,5 +100,6 @@ private: std::map kernel_map; std::map buffer_map; + }; diff --git a/include/RayCaster.h b/include/RayCaster.h index 65637e4..b9bc241 100644 --- a/include/RayCaster.h +++ b/include/RayCaster.h @@ -24,10 +24,12 @@ public: virtual void assign_map(Old_Map *map) = 0; virtual void assign_camera(Camera *camera) = 0; virtual void create_viewport(int width, int height, float v_fov, float h_fov) = 0; - virtual void assign_light(std::string light_id, Light light) = 0; + virtual void assign_lights(std::vector lights) = 0; + virtual void validate() = 0; // draw will abstract the gl sharing and software rendering // methods of retrieving the screen buffer + virtual void compute() = 0; virtual void draw(sf::RenderWindow* window) = 0; protected: @@ -35,11 +37,11 @@ protected: sf::Sprite viewport_sprite; sf::Texture viewport_texture; - Old_Map * map; - Camera *camera; - std::map light_map; - sf::Uint8 *viewport_image; - sf::Vector4f *viewport_matrix; + Old_Map * map = nullptr; + Camera *camera = nullptr; + std::vector lights; + sf::Uint8 *viewport_image = nullptr; + sf::Vector4f *viewport_matrix = nullptr; int error = 0; diff --git a/include/util.hpp b/include/util.hpp index 6ce8208..d51a35e 100644 --- a/include/util.hpp +++ b/include/util.hpp @@ -11,8 +11,13 @@ const float PI_F = 3.14159265358979f; struct Light { sf::Vector4f rgbi; + + // I believe that Vector3's get padded to Vector4's. Give them a non-garbage value sf::Vector3f position; + const float padding_1 = -1; + sf::Vector3f direction_cartesian; + const float padding_2 = -2; }; struct fps_counter { @@ -164,5 +169,6 @@ inline std::string read_file(std::string file_name){ std::stringstream buf; buf << input_file.rdbuf(); + input_file.close(); return buf.str(); } diff --git a/kernels/kernel.cl b/kernels/kernel.cl deleted file mode 100644 index 6acf0f8..0000000 --- a/kernels/kernel.cl +++ /dev/null @@ -1,125 +0,0 @@ -// global : local : constant : private - -// Function arguments of type image2d_t, image3d_t, image2d_array_t, image1d_t, image1d_buffer_t, -// and image1d_array_t refer to image memory objects allocated in the **global** address space. - -// http://downloads.ti.com/mctools/esd/docs/opencl/memory/buffers.html - -// Open CL C -// https://www.fixstars.com/en/opencl/book/OpenCLProgrammingBook/opencl-c/ - -__kernel void hello( - global int2* resolution, - global char* map, - global float3* projection_matrix, - global float3* cam_dir, - global float3* cam_pos, - global image2d_t* canvas) { - - printf("%s\n", "this is a test string\n"); - - - - const int MAX_RAY_STEPS = 64; - - // The pixel coord we are at - int2 screenPos = (int2)(get_global_id(0) % resolution->x, get_global_id(0) / resolution->x); - - // The X and Y planes - //float3 cameraPlaneU = vec3(1.0, 0.0, 0.0) - - // Y being multiplied by the aspect ratio, usually around .5-6ish; - //cl_float3 cameraPlaneV = vec3(0.0, 1.0, 0.0) * iResolution.y / iResolution.x; - - // So this is how they do that ray aiming! hah this is so tiny - // (camera direction) + (pixel.x * the X plane) + (product of pixel.y * Y plane) - // Oh all it's doing is adding the x and y coords of the pixel to the camera direction vector, interesting - - //cl_float3 rayDir = cameraDir + screenPos.x * cameraPlaneU + screenPos.y * cameraPlaneV; - - // the origin of the ray - // So the sign thing is for the up and down motion - - //cl_float3 rayPos = vec3(0.0, 2.0 * sin(iGlobalTime * 2.7), -12.0); - - // Ah, and here is where it spins around the center axis - // So it looks like its applying a function to rotate the x and z axis - //rayPos.xz = rotate2d(rayPos.xz, iGlobalTime); - //rayDir.xz = rotate2d(rayDir.xz, iGlobalTime); - - // Just an intvec of out coords - //ivec3 mapPos = ivec3(floor(rayPos)); - - // I think this is the delta t value - // the magnitude of the vector divided by the rays direction. Not sure what the aim of that is - // The ray direction might always be normalized, so that would be the dame as my delta_T - //vec3 deltaDist = abs(vec3(length(rayDir)) / rayDir); - - // The steps are the signs of the ray direction - //ivec3 rayStep = ivec3(sign(rayDir)); - - // ithe sign of the rays direction - // * - // Convert map position to a floating point vector and take away the ray position - // + - // the sign of the rays direction by 0.5 - // + - // 0.5 - // Now multyply everything by 0.5 - //vec3 sideDist = (sign(rayDir) * (vec3(mapPos) - rayPos) + (sign(rayDir) * 0.5) + 0.5) * deltaDist; - - // A byte mask - //bvec3 mask; - - // repeat until the max steps - //for (int i = 0; i < MAX_RAY_STEPS; i++) { - - // If there is a voxel at the map position, continue? - //if (getVoxel(mapPos)) - // break; - - // - // find which is smaller - // y ? z --> x` - // z ? x --> y` - // x ? y --> z` - // - // find which os is less or equal - // x` ? x --> x - // y` ? y --> y - // z` ? z --> z - - // Now find which ons is - //mask = lessThanEqual(sideDist.xyz, min(sideDist.yzx, sideDist.zxy)); - - - // Originally he used a component wise - /*bvec3 b1 = lessThan(sideDist.xyz, sideDist.yzx); - bvec3 b2 = lessThanEqual(sideDist.xyz, sideDist.zxy); - mask.x = b1.x && b2.x; - mask.y = b1.y && b2.y; - mask.z = b1.z && b2.z;*/ - //Would've done mask = b1 && b2 but the compiler is making me do it component wise. - - //All components of mask are false except for the corresponding largest component - //of sideDist, which is the axis along which the ray should be incremented. - - //sideDist += vec3(mask) * deltaDist; - //mapPos += ivec3(mask) * rayStep; - //} - - // Ah this is for coloring obviously, seems to be odd though, no indexing - //vec4 color; - //if (mask.x) { - // color = vec4(0.5); - //} - //if (mask.y) { - // color = vec4(1.0); - //} - //if (mask.z) { - // color = vec4(0.75); - //} - //write_imagef(image, pixel, color); - - -} \ No newline at end of file diff --git a/kernels/print_arguments.cl b/kernels/print_arguments.cl new file mode 100644 index 0000000..a62c09e --- /dev/null +++ b/kernels/print_arguments.cl @@ -0,0 +1,32 @@ +__kernel void printer( + global char* map, + global int3* map_dim, + global int2* resolution, + global float3* projection_matrix, + global float2* cam_dir, + global float3* cam_pos, + global float* lights, + global int* light_count, + __write_only image2d_t image +) { + + size_t id = get_global_id(0); + + if (id == 0) { + + printf("MAP: %i, %i, %i, %i", map[0], map[1], map[2], map[3]); + printf("MAP_DIMENSIONS: %i, %i, %i", map_dim[0].x, map_dim[0].y, map_dim[0].z); + printf("RESOLUTION: %i, %i", resolution[0].x, resolution[0].y); + printf("PROJECTION_MATRIX: %f, %f, %f", projection_matrix[0].x, projection_matrix[0].y, projection_matrix[0].z); + printf("CAMERA_DIRECTION: %f, %f", cam_dir[0].x, cam_dir[0].y); + printf("CAMERA_POSITION: %f, %f, %f", cam_pos[0].x, cam_pos[0].y, cam_pos[0].z); + printf("LIGHTS: %f, %f, %f, %f, %f, %f, %f, %f, %f, %f", lights[0], lights[1], lights[2], lights[3], lights[4], lights[5], lights[6], lights[7], lights[8], lights[9]); + printf("LIGHT_COUNT: %i", light_count); + + + + } + + return; + +} \ No newline at end of file diff --git a/kernels/ray_caster_kernel.cl b/kernels/ray_caster_kernel.cl index b082915..cb7a8ce 100644 --- a/kernels/ray_caster_kernel.cl +++ b/kernels/ray_caster_kernel.cl @@ -75,7 +75,7 @@ float4 cast_light_rays( // if it does } -__kernel void min_kern( +__kernel void raycaster( global char* map, global int3* map_dim, global int2* resolution, diff --git a/src/CL_Wrapper.cpp b/src/CL_Wrapper.cpp index 3790f3b..b4bb674 100644 --- a/src/CL_Wrapper.cpp +++ b/src/CL_Wrapper.cpp @@ -226,10 +226,11 @@ int CL_Wrapper::check_cl_khr_gl_sharing() { if (std::string(ext_str).find("cl_khr_gl_sharing") == std::string::npos) { std::cout << "No support for the cl_khr_gl_sharing extension"; - cl_khr_gl_sharing_fallback = true; + return -1; } delete ext_str; + return 1; } int CL_Wrapper::compile_kernel(std::string kernel_source, bool is_path, std::string kernel_name) { diff --git a/src/Curses.cpp b/src/Curses.cpp deleted file mode 100644 index 0528376..0000000 --- a/src/Curses.cpp +++ /dev/null @@ -1,187 +0,0 @@ -#pragma once -#include "Curses.h" -#include -#include - -Curses::Curses(sf::Vector2i tile_size_, sf::Vector2i grid_dimensions_) : - window(sf::VideoMode(tile_size_.x * grid_dimensions_.x, tile_size_.y * grid_dimensions_.y), "SimpleFML Curses"), - grid_dimensions(grid_dimensions_), - tile_pixel_dimensions(tile_size_){ - - font.loadFromFile("unifont.ttf"); - - for (int y = 0; y < grid_dimensions_.y; y++) { - for (int x = 0 ; x < grid_dimensions_.x; x++) { - tiles.emplace_back(Tile(sf::Vector2i(x, y))); - - // L'\u0020' = space char - } - } - - - // Set screen values to increasing unicode chars - /*wchar_t char_ind = L'\u0041'; - for (int i = 0; i < tiles.size(); i++) { - tiles.at(i).push_clear(char_ind++, sf::Color::White, sf::Color::Black); - }*/ -} - -Curses::~Curses() { -} - -void Curses::Update(double delta_time_) { - - for (Tile &tile : tiles) { - tile.inc_index(); - } - - sf::Event event; - while (window.pollEvent(event)) { - if (event.type == sf::Event::Closed) - window.close(); - } -} - -void Curses::Render() { - window.clear(); - - sf::Texture font_texture = font.getTexture(tile_pixel_dimensions.x); - - // Draw text and backfills - sf::VertexArray backfill_v_arr(sf::Quads, grid_dimensions.x * grid_dimensions.y * 4); - sf::VertexArray font_v_arr(sf::Quads, grid_dimensions.x * grid_dimensions.y * 4); - - int tile_index = 0; - - for (int i = 0; i < backfill_v_arr.getVertexCount(); i += 4) { - - Tile* tile = &tiles.at(tile_index); - sf::Glyph glyph = font.getGlyph(tile->current_unicode_value(), tile_pixel_dimensions.x, false); - - // Backfill the tile with the specified backfill color - backfill_v_arr[i + 0].position = sf::Vector2f(tile->getPosition().x * tile_pixel_dimensions.x, tile->getPosition().y * tile_pixel_dimensions.y); - backfill_v_arr[i + 1].position = sf::Vector2f(tile->getPosition().x * tile_pixel_dimensions.x + tile_pixel_dimensions.x, tile->getPosition().y * tile_pixel_dimensions.y); - backfill_v_arr[i + 2].position = sf::Vector2f(tile->getPosition().x * tile_pixel_dimensions.x + tile_pixel_dimensions.x, tile->getPosition().y * tile_pixel_dimensions.y + tile_pixel_dimensions.y); - backfill_v_arr[i + 3].position = sf::Vector2f(tile->getPosition().x * tile_pixel_dimensions.x, tile->getPosition().y * tile_pixel_dimensions.y + tile_pixel_dimensions.y); - - backfill_v_arr[i + 0].color = tile->current_backfill_color(); - backfill_v_arr[i + 1].color = tile->current_backfill_color(); - backfill_v_arr[i + 2].color = tile->current_backfill_color(); - backfill_v_arr[i + 3].color = tile->current_backfill_color(); - - // Draw the font with the correct size, texture location, window location, and color - font_v_arr[i + 0].position = sf::Vector2f(tile->getPosition().x * tile_pixel_dimensions.x, tile->getPosition().y * tile_pixel_dimensions.y); - font_v_arr[i + 1].position = sf::Vector2f(tile->getPosition().x * tile_pixel_dimensions.x + glyph.textureRect.width, tile->getPosition().y * tile_pixel_dimensions.y); - font_v_arr[i + 2].position = sf::Vector2f(tile->getPosition().x * tile_pixel_dimensions.x + glyph.textureRect.width, tile->getPosition().y * tile_pixel_dimensions.y + glyph.textureRect.height); - font_v_arr[i + 3].position = sf::Vector2f(tile->getPosition().x * tile_pixel_dimensions.x, tile->getPosition().y * tile_pixel_dimensions.y + glyph.textureRect.height); - - // Make the letter appear in the center of the tile by applying an offset - sf::Vector2f position_offset = sf::Vector2f(tile_pixel_dimensions.x / 4, tile_pixel_dimensions.y / 4); - font_v_arr[i + 0].position += position_offset; - font_v_arr[i + 1].position += position_offset; - font_v_arr[i + 2].position += position_offset; - font_v_arr[i + 3].position += position_offset; - - font_v_arr[i + 0].texCoords = sf::Vector2f(glyph.textureRect.left, glyph.textureRect.top); - font_v_arr[i + 1].texCoords = sf::Vector2f(glyph.textureRect.width + glyph.textureRect.left, glyph.textureRect.top); - font_v_arr[i + 2].texCoords = sf::Vector2f(glyph.textureRect.width + glyph.textureRect.left, glyph.textureRect.top + glyph.textureRect.height); - font_v_arr[i + 3].texCoords = sf::Vector2f(glyph.textureRect.left, glyph.textureRect.top + glyph.textureRect.height); - - font_v_arr[i + 0].color = tile->current_font_color(); - font_v_arr[i + 1].color = tile->current_font_color(); - font_v_arr[i + 2].color = tile->current_font_color(); - font_v_arr[i + 3].color = tile->current_font_color(); - - tile_index++; - } - - window.draw(backfill_v_arr); - window.draw(font_v_arr, &font_texture); - - window.display(); -} - -void Curses::setTile(Tile tile_) { - tiles.at(multi_to_linear(tile_.getPosition())) = tile_; -} - -void Curses::setTiles(std::vector tiles_) { - for (Tile tile: tiles_) { - tiles.at(multi_to_linear(tile.getPosition())) = tile; - } -} - -void Curses::Clear() { - for (Tile &tile : tiles) { - tile.clear(); - } -} - -Curses::Tile* Curses::getTile(sf::Vector2i position_) { - return &tiles.at(multi_to_linear(position_)); -} - -std::vector Curses::getTiles(sf::Vector2i start_, sf::Vector2i end_) { - std::vector ret; - for (int i = multi_to_linear(start_); i < multi_to_linear(end_); i++) { - ret.push_back(&tiles.at(i)); - } - return ret; -} - -void Curses::setScroll(int ratio_, sf::Vector2i start_, std::list scroll_text_) { - // Scrolling and it's scroll ratio is faux implemented by - // essentially stacking values and repeating them to slow the scroll down - - - - for (int i = 0; i < scroll_text_.size(); i++) { - append_slots(start_, scroll_text_); - scroll_text_.push_back(scroll_text_.front()); - scroll_text_.pop_front(); - } - - -} - - - -void Curses::set_tile_ratio(int ratio_, sf::Vector2i tile_position_) { - getTile(tile_position_)->set_ratio(ratio_); -} - -void Curses::append_slots(sf::Vector2i start_, std::list values_) -{ - std::vector tiles = getTiles(start_, - sf::Vector2i((values_.size() + start_.x) % grid_dimensions.x, - (values_.size() + start_.x) / grid_dimensions.x + start_.y)); - - if (tiles.size() != values_.size()) { - std::cout << "Values did not match up to slots when appending\n"; - return; - } - - std::list::iterator beg_slot_it = values_.begin(); - std::list::iterator end_slot_it = values_.end(); - std::list::iterator slot_it; - - std::vector::iterator beg_tile_it = tiles.begin(); - std::vector::iterator end_tile_it = tiles.end(); - std::vector::iterator tile_it; - - for (slot_it = beg_slot_it, tile_it = beg_tile_it; - (slot_it != end_slot_it) && (tile_it != end_tile_it); - ++slot_it, ++tile_it) { - - Tile* t = *tile_it; - t->push_back(*slot_it); - - } -} - -sf::Vector2i Curses::linear_to_multi(int position_) const { - return sf::Vector2i(position_ % grid_dimensions.x, position_ / grid_dimensions.x); -} -int Curses::multi_to_linear(sf::Vector2i position_) const { - return position_.y * grid_dimensions.x + position_.x; -} diff --git a/src/Hardware_Caster.cpp b/src/Hardware_Caster.cpp index 3a8fa4f..e63706e 100644 --- a/src/Hardware_Caster.cpp +++ b/src/Hardware_Caster.cpp @@ -30,7 +30,7 @@ int Hardware_Caster::init() { if (assert(error, "create_command_queue")) return error; - error = compile_kernel("../kernels/ray_caster_kernel.cl", true, "min_kern"); + error = compile_kernel("../kernels/ray_caster_kernel.cl", true, "raycaster"); if (assert(error, "compile_kernel")) { std::cin.get(); // hang the output window so we can read the error return error; @@ -45,8 +45,8 @@ void Hardware_Caster::assign_map(Old_Map *map) { this->map = map; auto dimensions = map->getDimensions(); - create_buffer("map_buffer", sizeof(char) * dimensions.x * dimensions.y * dimensions.z, map->get_voxel_data()); - create_buffer("dim_buffer", sizeof(int) * 3, &dimensions); + create_buffer("map", sizeof(char) * dimensions.x * dimensions.y * dimensions.z, map->get_voxel_data()); + create_buffer("map_dimensions", sizeof(int) * 3, &dimensions); } @@ -54,8 +54,43 @@ void Hardware_Caster::assign_camera(Camera *camera) { this->camera = camera; - create_buffer("cam_dir_buffer", sizeof(float) * 4, (void*)camera->get_direction_pointer(), CL_MEM_READ_ONLY | CL_MEM_USE_HOST_PTR); - create_buffer("cam_pos_buffer", sizeof(float) * 4, (void*)camera->get_position_pointer(), CL_MEM_READ_ONLY | CL_MEM_USE_HOST_PTR); + create_buffer("camera_direction", sizeof(float) * 4, (void*)camera->get_direction_pointer(), CL_MEM_READ_ONLY | CL_MEM_USE_HOST_PTR); + create_buffer("camera_position", sizeof(float) * 4, (void*)camera->get_position_pointer(), CL_MEM_READ_ONLY | CL_MEM_USE_HOST_PTR); +} + +void Hardware_Caster::validate() +{ + // Check to make sure everything has been entered; + if (camera == nullptr || + map == nullptr || + viewport_image == nullptr || + viewport_matrix == nullptr) { + + std::cout << "Raycaster.validate() failed, camera, map, or viewport not initialized"; + + } else { + + // Set all the kernel args + set_kernel_arg("raycaster", 0, "map"); + set_kernel_arg("raycaster", 1, "map_dimensions"); + set_kernel_arg("raycaster", 2, "viewport_resolution"); + set_kernel_arg("raycaster", 3, "viewport_matrix"); + set_kernel_arg("raycaster", 4, "camera_direction"); + set_kernel_arg("raycaster", 5, "camera_position"); + set_kernel_arg("raycaster", 6, "lights"); + set_kernel_arg("raycaster", 7, "light_count"); + set_kernel_arg("raycaster", 8, "image"); + + print_kernel_arguments(); + } + + +} + +void Hardware_Caster::compute() +{ + // correlating work size with texture size? good, bad? + run_kernel("raycaster", viewport_texture.getSize().x * viewport_texture.getSize().y); } // There is a possibility that I would want to move this over to be all inside it's own @@ -65,7 +100,7 @@ void Hardware_Caster::create_viewport(int width, int height, float v_fov, float // CL needs the screen resolution sf::Vector2i view_res(width, height); - create_buffer("res_buffer", sizeof(int) * 2, &view_res); + create_buffer("viewport_resolution", sizeof(int) * 2, &view_res); // And an array of vectors describing the way the "lens" of our // camera works @@ -110,7 +145,7 @@ void Hardware_Caster::create_viewport(int width, int height, float v_fov, float } } - create_buffer("view_matrix_buffer", sizeof(float) * 4 * view_res.x * view_res.y, viewport_matrix); + create_buffer("viewport_matrix", sizeof(float) * 4 * view_res.x * view_res.y, viewport_matrix); // Create the image that opencl's rays write to viewport_image = new sf::Uint8[width * height * 4]; @@ -129,16 +164,24 @@ void Hardware_Caster::create_viewport(int width, int height, float v_fov, float viewport_sprite.setTexture(viewport_texture); // Pass the buffer to opencl - create_image_buffer("image_buffer", sizeof(sf::Uint8) * width * height * 4, viewport_image); + create_image_buffer("image", sizeof(sf::Uint8) * width * height * 4, viewport_image); } -void Hardware_Caster::assign_light(std::string light_id, Light light) { +void Hardware_Caster::assign_lights(std::vector lights) { + + this->lights = std::vector(lights); + + int light_count = lights.size(); + + create_buffer("lights", sizeof(float) * 12 * light_count, lights.data(), CL_MEM_READ_ONLY | CL_MEM_USE_HOST_PTR); + + create_buffer("light_count", sizeof(int), &light_count); } void Hardware_Caster::draw(sf::RenderWindow* window) { - + window->draw(viewport_sprite); } int Hardware_Caster::acquire_platform_and_device() { @@ -286,8 +329,9 @@ int Hardware_Caster::create_shared_context() { int Hardware_Caster::create_command_queue() { + // If context and device_id have initialized if (context && device_id) { - // And the cl command queue + command_queue = clCreateCommandQueue(context, device_id, 0, &error); if (assert(error, "clCreateCommandQueue")) @@ -494,7 +538,7 @@ int Hardware_Caster::run_kernel(std::string kernel_name, const int work_size) { cl_kernel kernel = kernel_map.at(kernel_name); - error = clEnqueueAcquireGLObjects(getCommandQueue(), 1, &buffer_map.at("image_buffer"), 0, 0, 0); + error = clEnqueueAcquireGLObjects(getCommandQueue(), 1, &buffer_map.at("image"), 0, 0, 0); if (assert(error, "clEnqueueAcquireGLObjects")) return OPENCL_ERROR; @@ -510,13 +554,30 @@ int Hardware_Caster::run_kernel(std::string kernel_name, const int work_size) { clFinish(getCommandQueue()); // What if errors out and gl objects are never released? - error = clEnqueueReleaseGLObjects(getCommandQueue(), 1, &buffer_map.at("image_buffer"), 0, NULL, NULL); + error = clEnqueueReleaseGLObjects(getCommandQueue(), 1, &buffer_map.at("image"), 0, NULL, NULL); if (assert(error, "clEnqueueReleaseGLObjects")) return OPENCL_ERROR; return 1; } +void Hardware_Caster::print_kernel_arguments() +{ + compile_kernel("../kernels/print_arguments.cl", true, "printer"); + + set_kernel_arg("printer", 0, "map"); + set_kernel_arg("printer", 1, "map_dimensions"); + set_kernel_arg("printer", 2, "viewport_resolution"); + set_kernel_arg("printer", 3, "viewport_matrix"); + set_kernel_arg("printer", 4, "camera_direction"); + set_kernel_arg("printer", 5, "camera_position"); + set_kernel_arg("printer", 6, "lights"); + set_kernel_arg("printer", 7, "light_count"); + set_kernel_arg("printer", 8, "image"); + + run_kernel("printer", 1); +} + cl_device_id Hardware_Caster::getDeviceID() { return device_id; }; cl_platform_id Hardware_Caster::getPlatformID() { return platform_id; }; cl_context Hardware_Caster::getContext() { return context; }; diff --git a/src/RayCaster.cpp b/src/RayCaster.cpp index 8d48b62..e3a061d 100644 --- a/src/RayCaster.cpp +++ b/src/RayCaster.cpp @@ -1,110 +1,111 @@ #include "RayCaster.h" -#include -#include RayCaster::RayCaster() { } -void RayCaster::assign_map(Old_Map* map) { - this->map = map; -} - // Override values - //this.map_dimensions = new Vector3 (50, 50, 50); - //this.resolution = new Vector2 (200, 200); - //this.camera_direction = new Vector3 (1f, 0f, .8f); - //this.camera_position = new Vector3 (1, 10, 10); - - this->map_dimensions = map_dimensions; - this->map = map; - - resolution = viewport_resolution; - image = new sf::Color[resolution.x * resolution.y]; - - - // Calculate the view plane vectors - // Because casting to individual pixels causes barrel distortion, - // Get the radian increments - // Set the camera origin - // Rotate the ray by the specified pixel * increment - - double y_increment_radians = DegreesToRadians(50.0 / resolution.y); - double x_increment_radians = DegreesToRadians(80.0 / resolution.x); - - view_plane_vectors = new sf::Vector3f[resolution.x * resolution.y]; - for (int y = -resolution.y / 2 ; y < resolution.y / 2; y++) { - for (int x = -resolution.x / 2; x < resolution.x / 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 + resolution.x / 2) + resolution.x * (y + resolution.y / 2); - view_plane_vectors[index] = Normalize(ray); - } - } -} - RayCaster::~RayCaster() { - delete image; - delete view_plane_vectors; } -sf::Color* RayCaster::CastRays(sf::Vector3 camera_direction, sf::Vector3 camera_position) { - - // Setup the camera for this cast - this->camera_direction = camera_direction; - camera_direction_cartesian = Normalize(SphereToCart(camera_direction)); - this->camera_position = camera_position; - - // Start the loop at the top left, scan right and work down - for (int y = 0; y < resolution.y; y++) { - for (int x = 0; x < resolution.x; x++) { - - // Get the ray at the base direction - sf::Vector3f ray = view_plane_vectors[x + resolution.x * y]; - - // Rotate it to the correct pitch and yaw - - // Y axis, pitch - ray = sf::Vector3f( - ray.z * sin(camera_direction.y) + ray.x * cos(camera_direction.y), - ray.y, - ray.z * cos(camera_direction.y) - ray.x * sin(camera_direction.y) - ); - - // Z axis, yaw - ray = sf::Vector3f( - ray.x * cos(camera_direction.z) - ray.y * sin(camera_direction.z), - ray.x * sin(camera_direction.z) + ray.y * cos(camera_direction.z), - ray.z - ); - - - // Setup the ray - Ray r(map, resolution, sf::Vector2i(x, y), camera_position, ray); - - // Cast it and assign its return value - image[x + resolution.x * y] = r.Cast(); - } - } - - return image; -} - -void RayCaster::moveCamera(sf::Vector2f v) { - camera_direction.y += v.x; - camera_direction.z += v.y; -} +//void RayCaster::assign_map(Old_Map* map) { +// this->map = map; +//} +// // Override values +// //this.map_dimensions = new Vector3 (50, 50, 50); +// //this.resolution = new Vector2 (200, 200); +// //this.camera_direction = new Vector3 (1f, 0f, .8f); +// //this.camera_position = new Vector3 (1, 10, 10); +// +// this->map_dimensions = map_dimensions; +// this->map = map; +// +// resolution = viewport_resolution; +// image = new sf::Color[resolution.x * resolution.y]; +// +// +// // Calculate the view plane vectors +// // Because casting to individual pixels causes barrel distortion, +// // Get the radian increments +// // Set the camera origin +// // Rotate the ray by the specified pixel * increment +// +// double y_increment_radians = DegreesToRadians(50.0 / resolution.y); +// double x_increment_radians = DegreesToRadians(80.0 / resolution.x); +// +// view_plane_vectors = new sf::Vector3f[resolution.x * resolution.y]; +// for (int y = -resolution.y / 2 ; y < resolution.y / 2; y++) { +// for (int x = -resolution.x / 2; x < resolution.x / 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 + resolution.x / 2) + resolution.x * (y + resolution.y / 2); +// view_plane_vectors[index] = Normalize(ray); +// } +// } +//} +// +//RayCaster::~RayCaster() { +// delete image; +// delete view_plane_vectors; +//} +// +//sf::Color* RayCaster::CastRays(sf::Vector3 camera_direction, sf::Vector3 camera_position) { +// +// // Setup the camera for this cast +// this->camera_direction = camera_direction; +// camera_direction_cartesian = Normalize(SphereToCart(camera_direction)); +// this->camera_position = camera_position; +// +// // Start the loop at the top left, scan right and work down +// for (int y = 0; y < resolution.y; y++) { +// for (int x = 0; x < resolution.x; x++) { +// +// // Get the ray at the base direction +// sf::Vector3f ray = view_plane_vectors[x + resolution.x * y]; +// +// // Rotate it to the correct pitch and yaw +// +// // Y axis, pitch +// ray = sf::Vector3f( +// ray.z * sin(camera_direction.y) + ray.x * cos(camera_direction.y), +// ray.y, +// ray.z * cos(camera_direction.y) - ray.x * sin(camera_direction.y) +// ); +// +// // Z axis, yaw +// ray = sf::Vector3f( +// ray.x * cos(camera_direction.z) - ray.y * sin(camera_direction.z), +// ray.x * sin(camera_direction.z) + ray.y * cos(camera_direction.z), +// ray.z +// ); +// +// +// // Setup the ray +// Ray r(map, resolution, sf::Vector2i(x, y), camera_position, ray); +// +// // Cast it and assign its return value +// image[x + resolution.x * y] = r.Cast(); +// } +// } +// +// return image; +//} +// +//void RayCaster::moveCamera(sf::Vector2f v) { +// camera_direction.y += v.x; +// camera_direction.z += v.y; +//} diff --git a/src/main.cpp b/src/main.cpp index 139a65a..2393f73 100644 --- a/src/main.cpp +++ b/src/main.cpp @@ -7,7 +7,6 @@ #include #include #include -#include #include @@ -20,18 +19,16 @@ #endif +#pragma once #include #include #include #include #include - #include "Old_Map.h" -#include "Curses.h" #include "util.hpp" #include "RayCaster.h" #include "Hardware_Caster.h" -#include "CL_Wrapper.h" #include "Vector4.hpp" #include @@ -66,19 +63,11 @@ sf::Texture window_texture; int main() { - // It looks like I got the bulk of the stuff moved over to hardware caster. - // The lights still need work. Adding them to a map and checking for collisions - // will probably be the route I take. - // Need to hook up the assignment of kernel args - // Also need to hook up the rendering with the draw function. - - - sf::RenderWindow window(sf::VideoMode(WINDOW_X, WINDOW_Y), "SFML"); // Initialize the raycaster hardware, compat, or software RayCaster *rc = new Hardware_Caster(); - if (rc->init() != 0) { + if (rc->init() != 1) { delete rc; // rc = new Hardware_Caster_Compat(); // if (rc->init() != 0) { @@ -86,9 +75,8 @@ int main() { // rc = new Software_Caster(); // } } - - // This will be removed - CL_Wrapper c; + + // Set up the raycaster std::cout << "map..."; sf::Vector3i map_dim(MAP_X, MAP_Y, MAP_Z); @@ -98,92 +86,56 @@ int main() { rc->assign_map(map); Camera *camera = new Camera( - sf::Vector3f(0, 0, 0), - sf::Vector2f(0.0f, 1.00f) + sf::Vector3f(10, 11, 12), + sf::Vector2f(0.1f, 1.00f) ); rc->assign_camera(camera); rc->create_viewport(WINDOW_X, WINDOW_Y, 50.0f, 80.0f); - - int light_count = 2; - c.create_buffer("light_count_buffer", sizeof(int), &light_count); - - // {r, g, b, i, x, y, z, x', y', z'} - sf::Vector3f v = Normalize(sf::Vector3f(1.0f, 0.0f, 0.0f)); - sf::Vector3f v2 = Normalize(sf::Vector3f(1.1f, 0.4f, 0.7f)); - float light[] = { 0.4f, 0.8f, 0.1f, 1.0f, 50.0f, 50.0f, 50.0f, v.x, v.y, v.z, - 0.4f, 0.8f, 0.1f, 1.0f, 50.0f, 50.0f, 50.0f, v2.x, v2.y, v2.z}; - c.create_buffer("light_buffer", sizeof(float) * 10 * light_count, light, CL_MEM_READ_ONLY | CL_MEM_USE_HOST_PTR); - - - c.set_kernel_arg("min_kern", 0, "map_buffer"); - c.set_kernel_arg("min_kern", 1, "dim_buffer"); - c.set_kernel_arg("min_kern", 2, "res_buffer"); - c.set_kernel_arg("min_kern", 3, "view_matrix_buffer"); - c.set_kernel_arg("min_kern", 4, "cam_dir_buffer"); - c.set_kernel_arg("min_kern", 5, "cam_pos_buffer"); - c.set_kernel_arg("min_kern", 6, "light_buffer"); - c.set_kernel_arg("min_kern", 7, "light_count_buffer"); - c.set_kernel_arg("min_kern", 8, "image_buffer"); - - - // The step size in milliseconds between calls to Update() - // Lets set it to 16.6 milliseonds (60FPS) - float step_size = 0.0166f; - - // Timekeeping values for the loop - double frame_time = 0.0, - elapsed_time = 0.0, - delta_time = 0.0, - accumulator_time = 0.0, - current_time = 0.0; - fps_counter fps; + Light l; + l.direction_cartesian = sf::Vector3f(1.0f, 1.0f, 0.0f); + l.position = sf::Vector3f(10.0f, 10.0f, 10.0f); + l.rgbi = sf::Vector4f(0.3f, 0.4f, 0.3f, 1.0f); - // ============================= RAYCASTER SETUP ================================== + rc->assign_lights(std::vector{l}); - // Setup the sprite and texture - window_texture.create(WINDOW_X, WINDOW_Y); - window_sprite.setPosition(0, 0); + rc->validate(); - // State values + // Done setting up raycaster - sf::Vector3f cam_vec(0, 0, 0); - - //RayCaster ray_caster(map, map_dim, view_res); + // ========== DEBUG ========== + fps_counter fps; sf::Vector2f *dp = camera->get_direction_pointer(); debug_text cam_text_x(1, 30, &dp->x, "incli: "); debug_text cam_text_y(2, 30, &dp->y, "asmth: "); + // =========================== - sf::Vector3f *mp = camera->get_movement_pointer(); - debug_text cam_text_mov_x(4, 30, &mp->x, "X: "); - debug_text cam_text_mov_y(5, 30, &mp->y, "Y: "); - debug_text cam_text_mov_z(6, 30, &mp->y, "Z: "); - //debug_text cam_text_z(3, 30, &p->z); - debug_text light_x(7, 30, &light[7], "X: "); - debug_text light_y(8, 30, &light[8], "Y: "); - debug_text light_z(9, 30, &light[9], "Z: "); - // =============================================================================== + // 16.6 milliseconds (60FPS) + float step_size = 0.0166f; + double frame_time = 0.0, + elapsed_time = 0.0, + delta_time = 0.0, + accumulator_time = 0.0, + current_time = 0.0; // Mouse capture sf::Vector2i deltas; sf::Vector2i fixed(window.getSize()); bool mouse_enabled = true; - sf::Vector3f cam_mov_vec; - while (window.isOpen()) { // Poll for events from the user sf::Event event; while (window.pollEvent(event)) { - // If the user tries to exit the application via the GUI if (event.type == sf::Event::Closed) window.close(); + if (event.type == sf::Event::KeyPressed) { if (event.key.code == sf::Keyboard::Space) { if (mouse_enabled) @@ -194,10 +146,6 @@ int main() { } } - cam_vec.x = 0; - cam_vec.y = 0; - cam_vec.z = 0; - float speed = 1.0f; if (sf::Keyboard::isKeyPressed(sf::Keyboard::LShift)) { @@ -225,8 +173,6 @@ int main() { camera->set_position(sf::Vector3f(50, 50, 50)); } - camera->add_static_impulse(cam_vec); - if (mouse_enabled) { deltas = fixed - sf::Mouse::getPosition(); if (deltas != sf::Vector2i(0, 0) && mouse_enabled == true) { @@ -253,34 +199,14 @@ int main() { // ==== DELTA TIME LOCKED ==== } - float l[] = { - static_cast(light[9] * sin(delta_time / 1) + light[7] * cos(delta_time / 1)), - static_cast(light[8]), - static_cast(light[9] * cos(delta_time / 1) - light[7] * sin(delta_time / 1)) - }; - - float l2[] = { - static_cast(l[0] * cos(delta_time) - l[2] * sin(delta_time)), - static_cast(l[0] * sin(delta_time) + l[2] * cos(delta_time)), - static_cast(l[2]) - }; - - light[7] = l[0]; - light[8] = l[1]; - light[9] = l[2]; - // ==== FPS LOCKED ==== camera->update(delta_time); - // Run the raycast - rc->draw(&window); - //c.run_kernel("min_kern", WORK_SIZE); - - // ==== RENDER ==== - window.clear(sf::Color::Black); - window.draw(s); + // Run the raycast + rc->compute(); + rc->draw(&window); // Give the frame counter the frame time and draw the average frame time fps.frame(delta_time); @@ -288,14 +214,6 @@ int main() { cam_text_x.draw(&window); cam_text_y.draw(&window); - - cam_text_mov_x.draw(&window); - cam_text_mov_y.draw(&window); - cam_text_mov_z.draw(&window); - - light_x.draw(&window); - light_y.draw(&window); - light_z.draw(&window); window.display();