Commit d45e97f4 authored by unknown's avatar unknown
Browse files

Added missing shader files.

parent 26e57f29
#ifndef __GL_LINESPACE
#define __GL_LINESPACE
#include <raytracer/basic_structs.glh>
#include <raytracer/buffers.glh>
#include <raytracer/intersections.glh>
#include <raytracer/std.glh>
#include <raytracer/utilities.glh>
uint ls__faceRangeIndex(uint in_face, uint out_face)
{
return uint((((-0.5f * float(in_face)) + 4.5f)*float(in_face))) - 1u + out_face;
}
struct debug_data_t
{
uvec3 data_02;
uint data_01;
uvec4 data_03;
uvec4 data_05;
vec4 data_04;
};
STD_BUFFER debug_data_buffer
{
debug_data_t debug_data[];
};
// Equal to...
// uint offset = linespace.offsets[index];
//
//TODO: Wait for a fix. Or fix it somehow. Driver bug?
#define return_offset_at(i) case i: return linespace.offsets[i]
uint offsetof(const in Linespace linespace, const in uint index)
{
switch(index)
{
return_offset_at(0);
return_offset_at(1);
return_offset_at(2);
return_offset_at(3);
return_offset_at(4);
return_offset_at(5);
return_offset_at(6);
return_offset_at(7);
return_offset_at(8);
return_offset_at(9);
return_offset_at(10);
return_offset_at(11);
return_offset_at(12);
return_offset_at(13);
return_offset_at(14);
default: return 0;
}
}
uint getFaceWidth(uint axis){
switch(axis)
{
case 0:
case 1: return 2;
case 2: return 0;
}
}
uint getFaceHeight(uint axis){
switch(axis)
{
case 0:
case 2: return 1;
case 1: return 0;
}
}
uvec2 getPatch(const in Mesh mesh, vec3 at_position, int axis, int face){
bool flip_indices = face != axis;
return !flip_indices
? uvec2(at_position[getFaceWidth(axis)] / mesh.linespace.patch_size, at_position[getFaceHeight(axis)] / mesh.linespace.patch_size)
: uvec2(mesh.linespace.resolution[getFaceWidth(axis)], mesh.linespace.resolution[getFaceHeight(axis)]) -
uvec2(at_position[getFaceWidth(axis)] / mesh.linespace.patch_size, at_position[getFaceHeight(axis)] / mesh.linespace.patch_size)- 1;
}
float getDistance(const in Ray ray, const in Hit hit, inout uint dir_sign)
{
Mesh mesh = b_meshes[hit.mesh];
Triangle triangle = mesh.triangles[hit.triangle];
Vertex v1 = mesh.vertices[triangle.indices[1]];
Vertex v2 = mesh.vertices[triangle.indices[2]];
Vertex v3 = mesh.vertices[triangle.indices[0]];
vec3 vector = hit__getFromBarycentric(hit.barycentric, triangle, position).xyz - ray.origin;
dir_sign = uint(sign(dot(vector, ray.direction)));
return length(vector);
}
struct PatchInfo
{
int face;
int axis;
uvec2 patch_index;
vec3 position;
float t;
};
uint ls__lineIndex(const in Ray ray, const in Mesh mesh, float tmin, float tmax, int face_in, int face_out, inout bool direction_swapped)
{
Linespace linespace = mesh.linespace;
PatchInfo entry_info;
PatchInfo exit_info;
// face_in > face_out is an unsupported configuration for reasons of generation performance and memory conservation.
// If that happens, flip tmin, tmax and the two faces accordingly and mark it as flipped to we can take the farthest intersection in the shaft instead of the nearest.
direction_swapped = face_in > face_out;
// -------- t values -------------
entry_info.t = direction_swapped ? tmax : tmin;
exit_info.t = direction_swapped ? tmin : tmax;
// -------- faces ----------------
entry_info.face = direction_swapped ? face_out : face_in;
exit_info.face = direction_swapped ? face_in : face_out;
/////////////////////////////////////////////////////////////
//calculate entry and exit points on the minimum-normalized bounding box
//(so that bounds.min is at (0,0,0) and bounds.max is at the former bounding box's size)
entry_info.position = ray.origin + entry_info.t * ray.direction - linespace.bounds.min.xyz;
exit_info.position = ray.origin + exit_info.t * ray.direction - linespace.bounds.min.xyz;
entry_info.axis = entry_info.face % 3;
exit_info.axis = exit_info.face % 3;
// getPatch mirrors the patches on the negative axises per definition.
entry_info.patch_index = getPatch(mesh, entry_info.position, entry_info.axis, entry_info.face);
exit_info.patch_index = getPatch(mesh, exit_info.position, exit_info.axis, exit_info.face);
//Now with all that information, we can retrieve the line index.
uint offset = offsetof(linespace, ls__faceRangeIndex(entry_info.face, exit_info.face));
uint start_height = linespace.resolution[getFaceHeight(entry_info.axis)];
uint end_width = linespace.resolution[getFaceWidth(exit_info.axis)];
uint end_height = linespace.resolution[getFaceHeight(exit_info.axis)];
uint start = (entry_info.patch_index.y + entry_info.patch_index.x * start_height) * end_width * end_height;
uint end = (exit_info.patch_index.y + exit_info.patch_index.x * end_height);
return offset + start + end;
}
bool ls__traverse(const in Mesh mesh, const in Ray local_ray, const in bool use_first, const in float max_distance, inout Hit hit, inout float t_min)
{
float tmin;
float tmax;
int face_tmin;
int face_tmax;
Ray offset_ray = local_ray;
offset_ray.origin += local_ray.direction * 2*mesh.linespace.patch_size;
if(intersectsRayBounds(offset_ray, mesh.linespace.bounds, tmin, tmax, face_tmin, face_tmax))
{
if(t_min <= tmin || tmin > max_distance){
return false;
}
bool swapped = false;
uint line_id = ls__lineIndex(offset_ray, mesh, tmin, tmax, face_tmin, face_tmax, swapped);
Line line = mesh.lines[line_id];
uint dir_sign;
Hit nearer = swapped ? line.farthest : line.nearest;
nearer.mesh = mesh.id;
float nearest_distance = getDistance(offset_ray, nearer, dir_sign);
if(dir_sign > 0 && t_min > nearest_distance && max_distance > nearest_distance && nearer.barycentric.x != -1){
t_min = nearest_distance;
hit = nearer;
//Found an intersected shaft. But we don't yet know whether it contains geometry, so look it up as a last step.
return true;
}
nearer = swapped ? line.nearest : line.farthest;
nearer.mesh = mesh.id;
float farthest_distance = getDistance(offset_ray, nearer, dir_sign);
if(dir_sign > 0 && t_min > farthest_distance && max_distance > farthest_distance && nearer.barycentric.x != -1){
t_min = farthest_distance;
hit = nearer;
//Found an intersected shaft. But we don't yet know whether it contains geometry, so look it up as a last step.
return true;
}
}
return false;
}
#endif //__GL_LINESPACE
\ No newline at end of file
#ifndef GL_MESH_LS
#define GL_MESH_LS
#include <raytracer/basic_structs.glh>
#include <raytracer/buffers.glh>
#include <raytracer/intersections.glh>
#include <raytracer/std.glh>
#include <raytracer/utilities.glh>
const uvec3 c_face_width_axises = { 2, 2, 0 };
const uvec3 c_face_height_axises = { 1, 0, 1 };
uint faceRangeIndex(uint in_face, uint out_face)
{
return uint((((-0.5f * float(in_face)) + 4.5f)*float(in_face))) - 1u + out_face;
}
uvec2 getPatch(const in Mesh mesh, vec3 at_position, int axis, int face){
bool flip_indices = face>2;
return !flip_indices
? uvec2(at_position[c_face_width_axises[axis]] / mesh.linespace.patch_size, at_position[c_face_height_axises[axis]] / mesh.linespace.patch_size)
: uvec2(mesh.linespace.resolution[c_face_width_axises[axis]], mesh.linespace.resolution[c_face_height_axises[axis]]) -
uvec2(at_position[c_face_width_axises[axis]] / mesh.linespace.patch_size, at_position[c_face_height_axises[axis]] / mesh.linespace.patch_size);
}
struct PatchInfo
{
int face;
int axis;
uvec2 patch_index;
vec3 position;
float t;
};
float getDistance(const in Ray ray, const in Hit hit, inout float dir_sign)
{
Mesh mesh = b_meshes[hit.mesh];
Triangle triangle = mesh.triangles[hit.triangle];
Vertex v1 = mesh.vertices[triangle.indices[1]];
Vertex v2 = mesh.vertices[triangle.indices[2]];
Vertex v3 = mesh.vertices[triangle.indices[0]];
vec3 vector = hit__getFromBarycentric(hit.barycentric, triangle, position).xyz - ray.origin;
dir_sign = dot(vector, ray.direction);
return length(vector);
}
bool getLineIndex(const in Ray ray, const in Mesh mesh, inout uint line_id, inout bool direction_swapped)
{
float tmin, tmax;
int face_in, face_out;
direction_swapped = false;
//If the AABB is not intersected, cancel the traversal.
if (!intersectsRayBounds(ray, mesh.linespace.bounds, tmin, tmax, face_in, face_out))
return false;
PatchInfo entry_info;
PatchInfo exit_info;
// face_in > face_out is an unsupported configuration for reasons of generation performance and memory conservation.
// If that happens, flip tmin, tmax and the two faces accordingly and mark it as flipped to we can take the farthest intersection in the shaft instead of the nearest.
direction_swapped = face_in > face_out;
// -------- t values -------------
entry_info.t = direction_swapped ? tmax : tmin;
exit_info.t = direction_swapped ? tmin : tmax;
// -------- faces ----------------
entry_info.face = direction_swapped ? face_out : face_in;
exit_info.face = direction_swapped ? face_in : face_out;
/////////////////////////////////////////////////////////////
//calculate entry and exit points on the minimum-normalized bounding box
//(so that bounds.min is at (0,0,0) and bounds.max is at the former bounding box's size)
entry_info.position = ray.origin + entry_info.t * ray.direction - mesh.linespace.bounds.min.xyz;
exit_info.position = ray.origin + exit_info.t * ray.direction - mesh.linespace.bounds.min.xyz;
entry_info.axis = entry_info.face % 3;
exit_info.axis = exit_info.face % 3;
// getPatch mirrors the patches on the negative axises per definition.
entry_info.patch_index = getPatch(mesh, entry_info.position, entry_info.axis, entry_info.face);
exit_info.patch_index = getPatch(mesh, exit_info.position, exit_info.axis, exit_info.face);
//Now with all that information, we can retrieve the line index.
uint offset_index = faceRangeIndex(entry_info.face, exit_info.face);
uint offset = mesh.offsets[offset_index];
uint start_height = mesh.linespace.resolution[c_face_height_axises[entry_info.axis]];
uint end_width = mesh.linespace.resolution[c_face_width_axises[exit_info.axis]];
uint end_height = mesh.linespace.resolution[c_face_height_axises[exit_info.axis]];
uint start = (entry_info.patch_index.y + entry_info.patch_index.x * start_height ) * end_width * end_height;
uint end = (exit_info.patch_index.y + exit_info.patch_index.x * end_height);
//TODO: ERROR! I HAVE NO F*ING IDEA WHY BUT IT GOES HIGHER THAN LINE_COUNT!!!!!!!!!!!!!!!!!!!!!!!!!!!!!!!!!!!!!!!!!!
//Whatever.
line_id = clamp(offset + start + end, 0, 3839);//mesh.line_count-1
line_id = 0;
return true;
}
bool linespace__traverse(const in uint mesh_id, const in Ray ray, inout Hit hit, inout float distance) {
Mesh mesh = b_meshes[1];
//For ease of use, substitute Ray<->OBB for Ray<->AABB by transforming the ray into object space.
Ray ray_in = ray;
ray_in.direction = normalize(mat3(mesh.inverse_transformation) * ray_in.direction.xyz);
ray_in.origin = (mesh.inverse_transformation * vec4(ray_in.origin.xyz, 1)).xyz + 0.0*ray_in.direction;
hit = mesh.lines[0].nearest;
hit.mesh = mesh.id;
return true;
uint line_id;
bool direction_swapped;
if (getLineIndex(ray_in, mesh, line_id, direction_swapped))
{
Line line = mesh.lines[line_id];
float dir_sign;
float nearest_distance = getDistance(ray_in, direction_swapped ? line.farthest : line.nearest, dir_sign);
if(dir_sign > 0 && distance > nearest_distance){
hit = direction_swapped ? line.farthest : line.nearest;
hit.mesh = mesh.id;
distance = nearest_distance;
//Found an intersected shaft. But we don't yet know whether it contains geometry, so look it up as a last step.
return hit.barycentric.x != -1;
}
float farthest_distance = getDistance(ray_in, direction_swapped ? line.nearest : line.farthest, dir_sign);
if(dir_sign > 0 && distance > farthest_distance) {
hit = direction_swapped ? line.nearest : line.farthest;
hit.mesh = mesh.id;
distance = farthest_distance;
//Found an intersected shaft. But we don't yet know whether it contains geometry, so look it up as a last step.
return hit.barycentric.x != -1;
}
}
return false;
}
#endif //GL_MESH_LS
\ No newline at end of file
#include <glare_core>
#include <glare_advanced>
#include <glare_io>
using namespace glare;
int main(int argc, char* argv[])
{
io::Arguments arguments(argc, argv);
std::string mask_asset_path = arguments.get("o", 0, std::string(""));
advanced::LineSpaceMaskMode mask_mode = arguments.has("ns") ? advanced::LineSpaceMaskMode::eNonSymmetrical : advanced::LineSpaceMaskMode::eSymmetrical;
unsigned mask_resolution_min = arguments.get("r", 0, 1);
unsigned mask_resolution_max = arguments.get("r", 1, 16);
bool check_after = arguments.get("check", 0, false);
advanced::LineSpaceMaskGenerator::generate(mask_mode, math::Range<unsigned>(mask_resolution_min, mask_resolution_max), mask_asset_path, check_after);
quitPromptDefault(0);
return 0;
}
\ No newline at end of file
#include "BVH.h"
#include <engine/rendering/Shader.h>
#include <engine/EngineState.h>
#include <engine/Math.h>
namespace glare
{
namespace advanced
{
BVH::BVH()
{
}
void BVH::initialize(const std::shared_ptr<SceneCollector> &collector)
{
m_scene_collector = collector;
m_mesh_index_offset = 0;
build();
buildBuffers();
}
BVH::~BVH()
{
}
std::shared_ptr<core::Buffer<gl::BufferType::eShaderStorage>> BVH::getTrianglesBuffer() const
{
return m_triangles_buffer;
}
std::shared_ptr<core::Buffer<gl::BufferType::eShaderStorage>> BVH::getBVHNodesBuffer() const
{
return m_bvh_nodes_buffer;
}
void BVH::buildBuffers()
{
if (!m_triangles_buffer)
m_triangles_buffer = std::make_shared<core::Buffer<gl::BufferType::eShaderStorage>>();
if (!m_bvh_nodes_buffer)
m_bvh_nodes_buffer = std::make_shared<core::Buffer<gl::BufferType::eShaderStorage>>();
m_triangles_buffer->setData(m_sorted_triangles, gl::BufferUsage::eDynamicRead);
m_bvh_nodes_buffer->setData(m_nodes, gl::BufferUsage::eDynamicRead);
}
void BVH::build()
{
m_sorted_triangles = std::vector<Triangle>(m_scene_collector->triangles());
m_nodes.resize(2 * m_sorted_triangles.size());
m_triangle_centroids.resize(m_sorted_triangles.size());
m_centroid_boxes.resize(m_sorted_triangles.size());
m_safe_centroid_boxes.resize(m_sorted_triangles.size());
m_ranges.clear();
m_node_counter = 0;
math::Bounds centroid_box;
#pragma omp parallel for schedule(static) //static, because workload is the same for every iteration
for (int i = 0; i < int(m_sorted_triangles.size()); ++i)
{
// [] is not necessarily range-checked, whereas at() has to be range checked by standard
const Triangle& triangle = m_sorted_triangles[i];
//calculate triangle centroid
glm::vec4 centroid = (m_scene_collector->vertices()[triangle.indices[0]].position
+ m_scene_collector->vertices()[triangle.indices[1]].position
+ m_scene_collector->vertices()[triangle.indices[2]].position)
/ 3.f;
m_triangle_centroids[i] = centroid;
centroid_box.expand(centroid);
}
m_centroid_boxes[0] = centroid_box;
TriangleIndexRange range;
range.start = 0;
range.end = int(m_sorted_triangles.size() - 1);
range.parent = -1;
m_ranges.push_back(range);
m_ranges.resize(m_sorted_triangles.size());
m_safe_ranges.resize(m_sorted_triangles.size());
m_kernel_counter = 1;
m_bvh_depth = 0;
while (m_kernel_counter > 0)
{
m_bvh_depth++;
m_buffer_position = 0;
const int kernel_counter = m_kernel_counter;
#pragma omp parallel for schedule(dynamic) //dynamic, because construct takes different time dynamically
for (int global_id = 0; global_id < kernel_counter; global_id++)
{
split(global_id);
}
m_node_counter += kernel_counter;
//secure reads
if (m_kernel_counter > 0)
{
#pragma omp parallel for schedule(static) //static, because workload is the same for every iteration
for (int global_id = 0; global_id < m_kernel_counter; global_id++)
{
std::swap(m_ranges[global_id], m_safe_ranges[global_id]);
std::swap(m_centroid_boxes[global_id], m_safe_centroid_boxes[global_id]);
}
}
}
m_centroid_boxes.clear();
m_safe_centroid_boxes.clear();
m_ranges.clear();
m_safe_ranges.clear();
m_triangle_centroids.clear();
}
bool BVH::shouldSplit(const math::Bounds &centroid)
{
glm::vec4 diff = centroid[1] - centroid[0];
return diff.x > 0.f && diff.y > 0.f && diff.z > 0.f;
}
void BVH::addAxisAlignedBoundingBox(BVHNode& node, int start, int end) const
{
glm::vec4 best_min = glm::vec4(std::numeric_limits<float>::max(), std::numeric_limits<float>::max(), std::numeric_limits<float>::max(), 1.0f);
glm::vec4 best_max = glm::vec4(std::numeric_limits<float>::lowest(), std::numeric_limits<float>::lowest(), std::numeric_limits<float>::lowest(), 1.0f);
for (int id = start; id <= end; id++)
{
Triangle t = m_sorted_triangles.at(id);
glm::vec4 a = m_scene_collector->vertices().at(t.indices[0]).position;
glm::vec4 b = m_scene_collector->vertices().at(t.indices[1]).position;
glm::vec4 c = m_scene_collector->vertices().at(t.indices[2]).position;
glm::vec4 current_min = glm::min(glm::min(a, b), c);
glm::vec4 current_max = glm::max(glm::max(a, b), c);
best_min = glm::min(current_min, best_min);
best_max = glm::max(current_max, best_max);
}
node.bounds.min = best_min;
node.bounds.max = best_max;
}
void BVH::split(uint32_t global_id)
{
static const float flt_max = std::numeric_limits<float>::max();
static const float flt_min = std::numeric_limits<float>::lowest();
int start = m_ranges[global_id].start;
int end = m_ranges[global_id].end;
int parent = m_ranges[global_id].parent;
//centboxes cover the centroids of the triangles, not the whole objects
math::Bounds cebo = m_centroid_boxes[global_id];
bool splitting = parent < 0 || shouldSplit(cebo);
// since the thread will now be processed the value of the kernel_counter is decreased
--m_kernel_counter;
int leaf_threshold = 1;
BVHNode node;
// current_node_id is the position in the node buffer in which the new node is stored
int current_node_id = m_node_counter + global_id;
int diff = end - start;
if ((diff > leaf_threshold) && splitting)
{ //inner node
//pos is the position in which the new ranges will be stored
int pos = m_buffer_position.fetch_add(1);
TriangleIndexRange left_range, right_range;
math::Bounds left_centbox, right_centbox;
// since the range is divided in two parts more threads can process the construction.
// therefore the kernel_counter is increased by 2
m_kernel_counter.fetch_add(2);
// 1. the AABB which contains the range's triangle information is added for later processing (traversal)
addAxisAlignedBoundingBox(node, start, end);
// 0 is the flag for inner node in the node buffer
node.type = 0;
// 2. the split plane which causes the least cost is determined by iterating over the
// 3 axes (since it's axis aligned)
float best = flt_max;
int best_axis = 0;
int best_plane = 0;
AxisConstants axis_constants[3];
Bin bins[3][16];
for (int axis = 0; axis < 3; ++axis)
{
axis_constants[axis].cbmin = cebo[0][axis];
axis_constants[axis].cbmax = cebo[1][axis];
axis_constants[axis].cbdiff = axis_constants[axis].cbmax - axis_constants[axis].cbmin;
axis_constants[axis].k = (m_bin_count * (1 - m_bin_epsilon)) / axis_constants[axis].cbdiff;
//Initialize Bins with default values.
for (int32_t i = 0; i < m_bin_count; ++i)
{
bins[axis][i].bounds.max = glm::vec4(flt_min);
bins[axis][i].bounds.min = glm::vec4(flt_max);
bins[axis][i].object_count = 0;
}
}
// This is the thing that takes too damn long
for (int id = start; id <= end; ++id)
{
glm::vec4 centroidAxis = m_triangle_centroids[id];
for (int axis = 0; axis < 3; ++axis)
{
int bin_id = int(axis_constants[axis].k * (centroidAxis[axis] - axis_constants[axis].cbmin));
bins[axis][bin_id].bounds.expand(centroidAxis);
++(bins[axis][bin_id].object_count);
}
}
for (int axis = 0; axis < 3; ++axis)
{
Bin left_bounds[15];
//Again, initialize all bounds.
for (int i = 0; i < m_plane_count; ++i)
{
left_bounds[i].bounds.max = glm::vec4(flt_min);
left_bounds[i].bounds.min = glm::vec4(flt_max);
left_bounds[i].object_count = 0;
}
left_bounds[0].bounds.expand(bins[axis][0].bounds);
left_bounds[0].object_count = bins[axis][0].object_count;
for (int plane = 1; plane < m_plane_count; ++plane)
{
left_bounds[plane].bounds.expand(left_bounds[plane - 1].bounds);
left_bounds[plane].bounds.expand(bins[axis][plane].bounds);
left_bounds[plane].object_count = left_bounds[plane - 1].object_count + bins[axis][plane].object_count;
}
Bin right_bounds[15];
//Again and again, initialize bounds.
for (int i = 0; i < m_plane_count; ++i)
{
right_bounds[i].bounds.max = glm::vec4(flt_min);
right_bounds[i].bounds.min = glm::vec4(flt_max);
right_bounds[i].object_count = 0;
}
for (int plane = m_plane_count - 1; plane >= 0; --plane)
{
right_bounds[plane].bounds.expand(bins[axis][plane + 1].bounds);
right_bounds[plane].object_count = bins[axis][plane + 1].object_count;
if (plane != m_plane_count - 1)
{
right_bounds[plane].bounds.expand(right_bounds[plane + 1].bounds);