From 2e6ddd2ff23467d673678d0cf05e2989c577c8b2 Mon Sep 17 00:00:00 2001 From: mo7sener Date: Wed, 25 Dec 2019 00:09:14 +0200 Subject: [PATCH] Adding basic texture fragment shader kernel --- OpenWindow/OpenWindow.vcxproj | 8 +- OpenWindow/camera.cpp | 21 ++-- OpenWindow/kernel_sources.cpp | 112 ++++++++++++++++++++++ OpenWindow/kernels.cpp | 176 ++++++++++++++++++++++++++++++---- OpenWindow/kernels.h | 23 +++-- OpenWindow/main.cpp | 3 + OpenWindow/model.h | 4 +- OpenWindow/renderer.cpp | 106 +++++++++++++++----- OpenWindow/tgaimage.h | 3 +- OpenWindow/util_renderer.cpp | 41 +++++--- OpenWindow/util_window.cpp | 4 +- 11 files changed, 412 insertions(+), 89 deletions(-) diff --git a/OpenWindow/OpenWindow.vcxproj b/OpenWindow/OpenWindow.vcxproj index d6e46ec..7b45e66 100644 --- a/OpenWindow/OpenWindow.vcxproj +++ b/OpenWindow/OpenWindow.vcxproj @@ -76,7 +76,7 @@ Disabled true true - $(ProjectDir)\include + $(ProjectDir)\include;%(AdditionalIncludeDirectories) 4996; _MBCS;%(PreprocessorDefinitions) @@ -91,7 +91,7 @@ Disabled true true - $(ProjectDir)\include + $(ProjectDir)\include;%(AdditionalIncludeDirectories) 4996; _MBCS;%(PreprocessorDefinitions) @@ -108,7 +108,7 @@ true true true - $(ProjectDir)\include + $(ProjectDir)\include;%(AdditionalIncludeDirectories) 4996; _MBCS;%(PreprocessorDefinitions) @@ -127,7 +127,7 @@ true true true - $(ProjectDir)\include + $(ProjectDir)\include;%(AdditionalIncludeDirectories) 4996; _MBCS;%(PreprocessorDefinitions) diff --git a/OpenWindow/camera.cpp b/OpenWindow/camera.cpp index 57af509..5a71ccd 100644 --- a/OpenWindow/camera.cpp +++ b/OpenWindow/camera.cpp @@ -101,31 +101,24 @@ Matrix Camera::GetModelViewMatrix() { Vec3f center = position + forward; Vec3f z = forward * -1; Matrix Minv = Matrix::identity(); - Matrix Tr = Matrix::identity(); for (int i = 0; i < 3; i++) { Minv[0][i] = right[i]; Minv[1][i] = up[i]; Minv[2][i] = z[i]; - Tr[i][3] = -center[i]; } -// Minv[3][0] = (right * position); -// Minv[3][1] = (up * position); -// Minv[3][2] = (z * position); - return Minv * Tr; + Minv[0][3] = -(right * position); + Minv[1][3] = -(up * position); + Minv[2][3] = -(z * position); + return Minv; } Matrix Camera::GetProjectionMatrix() { Matrix Projection = Matrix::identity(); - Projection[0][0] = 1 / tan(fov * DEG2RAD); - Projection[1][1] = 1 / tan(fov * DEG2RAD); + Projection[0][0] = 1 / tan(fov * DEG2RAD / 2); + Projection[1][1] = 1 / tan(fov * DEG2RAD / 2); Projection[2][2] = (far_plane + near_plane) / (far_plane - near_plane); Projection[2][3] = (2 * far_plane * near_plane) / (far_plane - near_plane); - //Projection[2][2] = -(far_plane) / (far_plane - near_plane); - //Projection[2][3] = -1; - //Projection[3][2] = -(far_plane * near_plane) / (far_plane - near_plane); - //Projection[2][2] = -2 / (far_plane - near_plane); - //Projection[2][3] = (far_plane + near_plane) / (far_plane - near_plane); Projection[3][2] = -1; - Projection[3][3] = 1; + Projection[3][3] = 0; return Projection; } diff --git a/OpenWindow/kernel_sources.cpp b/OpenWindow/kernel_sources.cpp index f1e480a..494ffc3 100644 --- a/OpenWindow/kernel_sources.cpp +++ b/OpenWindow/kernel_sources.cpp @@ -14,3 +14,115 @@ const char* vertex_shader_kernel_source = " + m[local_index*4 + 2]*VertexBuffer[3*global_index+2] \n" " + m[local_index*4 + 3]; \n" "} \n"; + +const char* fragment_shader_kernel_source = +"float3 barycentric(float3* pts, float3 P) \n" +"{ \n" +" float3 u = cross( \n" +" (float3){pts[0][2] - pts[0][0], pts[0][1] - pts[0][0], pts[0][0] - P[0]}, // AC_x, AB_x, distance_x \n" +" (float3){pts[1][2] - pts[1][0], pts[1][1] - pts[1][0], pts[1][0] - P[1]} // AC_y, AB_y, distance_y \n" +" ); \n" +" if (fabs(u[2]) < 1) return (float3){-1, 1, 1}; \n" +" return (float3){1.f - (u[0] + u[1]) / u[2], u[1] / u[2], u[0] / u[2]}; \n" +"} \n" +" \n" +"__kernel void fragment_shader ( \n" +" __global int* faces, \n" +" __global float* vertices, \n" +" __global int* pixels, \n" +" __global int* screen_width, \n" +" __global float* z_buffer, \n" +" __global int* nfaces, \n" +" __global float* uv_buffer, \n" +" __global int* map_size, \n" +" __global float* light_dir, \n" +" __global float* norms_buff, \n" +" __global uchar* diffuse_map \n" +") { \n" +" int index = get_global_id(0); \n" +" \n" +" \n" +" bool out = true; \n" +" float3 vertices3[3]; \n" +" float2 uv_coords[3]; \n" +" float3 norms[3]; \n" +" \n" +" for(int i = 0; i < 3; i++) { \n" +" float4 vertex = vertices [faces[(index * 9) + (3 * i) + (0)]]; \n" +" for(int j = 0; j < 4; j ++) { \n" +" vertex[j] = vertices[4 * faces[(index*9) + (3 * i) + (0)] + j]; \n" +" } \n" +" \n" +" for(int j = 0; j < 2; j++) { \n" +" uv_coords[i][j] = uv_buffer [2 * faces[(index * 9) + (3 * i) + (1)] + j]; \n" +" } \n" +" \n" +" for( int j = 0; j < 3; j++ ) { \n" +" vertices3[j][i] = (vertex[j]/vertex[3]); \n" +" norms[j][i] = norms_buff[3 * (faces[(index * 9) + (3 * i) + (2)]) + j]; \n" +" } \n" +" \n" +" if ( vertices3[0][i] > 0 && vertices3[0][i] < *screen_width \n" +" && vertices3[1][i] > 0 && vertices3[1][i] < *screen_width ) \n" +" out = false; \n" +" } \n" +" \n" +" if(out) return; \n" +" \n" +" //if(vertices3[1][0] == vertices3[1][1] && vertices3[1][2] == vertices3[1][1]) return; \n" +" \n" +" int2 bounding_box_min = (int2) { *screen_width - 1, *screen_width - 1 }; \n" +" int2 bounding_box_max = (int2) { 0, 0 }; \n" +" int2 clamper = (int2) { *screen_width - 1, *screen_width - 1 }; \n" +" \n" +" for(int i = 0; i < 3; i++) { \n" +" for(int j = 0; j < 2; j++) { \n" +" bounding_box_min[j] = max(0, min(bounding_box_min[j], (int)vertices3[j][i])); \n" +" bounding_box_max[j] = min(clamper[j], max(bounding_box_max[j], (int)vertices3[j][i])); \n" +" } \n" +" } \n" +" \n" +" float3 point; \n" +" for(point[0] = bounding_box_min[0]; point[0] <= bounding_box_max[0]; point[0]++) { \n" +" for(point[1] = bounding_box_min[1]; point[1] <= bounding_box_max[1]; point[1]++) { \n" +" \n" +" float3 bc_coord = barycentric(vertices3, point); \n" +" if (bc_coord[0] < 0 || bc_coord[1] < 0 || bc_coord[2] < 0) continue; \n" +" \n" +" float2 uv_vec = (float2){0, 0}; \n" +" float3 normal = (float3){0, 0, 0}; \n" +" \n" +" point[2] = dot(vertices3[2], bc_coord); \n" +" if (z_buffer[(int)(point[0] + point[1] * *screen_width)] > point[2]) { \n" +" continue; \n" +" } \n" +" \n" +" \n" +" for (int i = 0; i < 3; i++) { \n" +" uv_vec[0] += uv_coords[i][0] * bc_coord[i]; \n" +" uv_vec[1] += uv_coords[i][1] * bc_coord[i]; \n" +" normal[i] = dot(norms[i], bc_coord); \n" +" } \n" +" int2 uv_point = (int2) { (int)(uv_vec[0] * map_size[0]), (int)(uv_vec[1] * map_size[1]) }; \n" +" \n" +" \n" +" int col_index = 3 * (uv_point[0] + uv_point[1] * map_size[0]); \n" +" \n" +" \n" +" \n" +" \n" +" float3 normalized_norm = normalize(normal); \n" +" \n" +" \n" +" float intensity = clamp((dot(normalized_norm , (float3){light_dir[0], light_dir[1], light_dir[2]})), 0.f, 1.f); \n" +" \n" +" int color = 0; \n" +" color |= ((int)fmin((float)(diffuse_map[col_index + 0]) * intensity, (float) 0xff)) << 16; \n" +" color |= ((int)fmin((float)(diffuse_map[col_index + 1]) * intensity, (float) 0xff)) << 8; \n" +" color |= ((int)fmin((float)(diffuse_map[col_index + 2]) * intensity, (float) 0xff)) << 0; \n" +" \n" +" z_buffer[(int) (point[0] + point[1] * *screen_width)] = point[2]; \n" +" pixels [(int) (point[0] + point[1] * *screen_width)] = color;// & 0x00ffffff; \n" +" } \n" +" } \n" +"} \n"; diff --git a/OpenWindow/kernels.cpp b/OpenWindow/kernels.cpp index c2e3f48..5f31cf7 100644 --- a/OpenWindow/kernels.cpp +++ b/OpenWindow/kernels.cpp @@ -1,5 +1,8 @@ #include "kernels.h" +#include "util_window.h" #include +#include +#include cl_uint numPlatforms; cl_uint numDevices; @@ -8,7 +11,7 @@ cl_platform_id platform_id; cl_device_id device_id; cl_context context; - +int triangles_drawn = 0; // Matrix Multiplication Variables #define MATRIX_SIZE 16 @@ -17,13 +20,49 @@ cl_context context; cl_mem mat_z; cl_mem vertices_mem; cl_mem new_vertices_mem; +bool vertex_shader_buffers_initialized = false; +//================== +// fragment shader +// int* faces, +// int nfaces, +// float* new_verts, +// int nverts, +// int screen_width, +// int screen_height, +// float* z_buffer, +// float* uniform_m, +// float* uniform_mit, +// float* light_dir, +// const char* diffuse_map, +// const char* normal_map, +// const char* spec_map, +// char* pixel_data +cl_mem faces_buffer; +cl_mem uniform_m; +cl_mem uniform_mit; +cl_mem light_dir_buffer; +cl_mem diffuse_map_buffer; +cl_mem norms_mem; +cl_mem spec_map_buffer; +int* hidden_pixel_buffer; +bool fragment_shader_buffers_initialized = false; +cl_mem triangles_verts_mem; +cl_mem bounding_box_min_mem; +cl_mem z_buffer_mem; +cl_mem pixel_data_buffer; +cl_mem screen_width_mem; +cl_mem nfaces_mem; +cl_mem uv_buffer; +cl_mem map_size_buffer; //================== cl_platform_id* platforms = NULL; cl_device_id* devices = NULL; cl_program vertex_shader_prog; +cl_program fragment_shader_prog; cl_kernel vertex_shader_kernel; +cl_kernel fragment_shader_kernel; cl_command_queue commands; int err; @@ -44,58 +83,153 @@ void init_kernels() { commands = clCreateCommandQueue(context, devices[0], 0, &err); vertex_shader_prog = clCreateProgramWithSource(context, 1, (const char **)&vertex_shader_kernel_source, NULL, &err); + fragment_shader_prog = clCreateProgramWithSource(context, 1, (const char**)&fragment_shader_kernel_source, NULL, &err); err = clBuildProgram(vertex_shader_prog, 1, devices, NULL, NULL, NULL); + err = clBuildProgram(fragment_shader_prog, 1, devices, NULL, NULL, NULL); if (err != CL_SUCCESS) { size_t len; char buffer[2048]; - printf("Error: Failed to build program executable!\n"); - //clGetProgramBuildInfo(mat_mul_prog, device_id, CL_PROGRAM_BUILD_LOG, sizeof(buffer), buffer, &len); + printf("Error: Failed to build the fragment shader prog!\n"); + clGetProgramBuildInfo(fragment_shader_prog, devices[0], CL_PROGRAM_BUILD_LOG, sizeof(buffer), buffer, &len); printf(buffer); - return; + //return; } vertex_shader_kernel = clCreateKernel(vertex_shader_prog, "vertex_shader", &err); + fragment_shader_kernel = clCreateKernel(fragment_shader_prog, "fragment_shader", &err); err = 0; } void vertex_shader(float* z, float* vertices, int vertex_count, float* new_vertices) { - mat_z = clCreateBuffer(context, CL_MEM_READ_ONLY , sizeof(float) * MATRIX_SIZE , NULL, &err); - vertices_mem = clCreateBuffer(context, CL_MEM_READ_ONLY , sizeof(float) * vertex_count * 3, NULL, &err); - new_vertices_mem = clCreateBuffer(context, CL_MEM_WRITE_ONLY, sizeof(float) * vertex_count * 4, NULL, &err); + if(!vertex_shader_buffers_initialized) + { + mat_z = clCreateBuffer(context, CL_MEM_READ_ONLY , sizeof(float) * MATRIX_SIZE , NULL, &err); + vertices_mem = clCreateBuffer(context, CL_MEM_READ_ONLY , sizeof(float) * vertex_count * 3, NULL, &err); + new_vertices_mem = clCreateBuffer(context, CL_MEM_WRITE_ONLY, sizeof(float) * vertex_count * 4, NULL, &err); - err = clSetKernelArg(vertex_shader_kernel, 0, sizeof(cl_mem), &mat_z); - err = clSetKernelArg(vertex_shader_kernel, 1, sizeof(cl_mem), &vertices_mem); - err = clSetKernelArg(vertex_shader_kernel, 2, sizeof(cl_mem), &new_vertices_mem); + err = clSetKernelArg(vertex_shader_kernel, 0, sizeof(cl_mem), &mat_z); + err = clSetKernelArg(vertex_shader_kernel, 1, sizeof(cl_mem), &vertices_mem); + err = clSetKernelArg(vertex_shader_kernel, 2, sizeof(cl_mem), &new_vertices_mem); + vertex_shader_buffers_initialized = true; + } err = clEnqueueWriteBuffer(commands, mat_z , CL_TRUE, 0, sizeof(float) * MATRIX_SIZE, z, 0, NULL, NULL); err = clEnqueueWriteBuffer(commands, vertices_mem, CL_TRUE, 0, sizeof(float) * vertex_count * 3, vertices, 0, NULL, NULL); size_t vertex_shader_global[] = { vertex_count*4 }; size_t vertex_shader_local[] = { 4 }; - err = clEnqueueNDRangeKernel(commands, vertex_shader_kernel, 1, NULL, vertex_shader_global, vertex_shader_local, 0, NULL, NULL); +} - clFlush(commands); - clFinish(commands); - err = clEnqueueReadBuffer(commands, new_vertices_mem, CL_TRUE, 0, sizeof(float) * vertex_count * 4, new_vertices, 0, NULL, NULL); +void clear_pixel_buffer() { + for (int i = 0; i < screen_width * screen_height; i++) { + hidden_pixel_buffer[i] = 0; + } +} + +void clear(cl_mem* buffer, size_t size, const int pattern) { + clEnqueueFillBuffer(commands, *buffer, &pattern, sizeof(int), 0, size, 0, NULL, NULL); +} + +void fragment_shader( + int* faces, + int nfaces, + float* uv, + size_t uv_size, + float* uniform_m, + float* uniform_mit, + float* light_dir, + unsigned char* diffuse_map, + float* norms, + size_t norms_size, + unsigned char* spec_map, + int* map_size +) { + if (!fragment_shader_buffers_initialized) { + + faces_buffer = clCreateBuffer(context, CL_MEM_READ_ONLY , sizeof(int) * 3 * 3 * nfaces , NULL, &err); + pixel_data_buffer = clCreateBuffer(context, CL_MEM_WRITE_ONLY, sizeof(int) * screen_height * screen_width , NULL, &err); + screen_width_mem = clCreateBuffer(context, CL_MEM_READ_ONLY , sizeof(int) , NULL, &err); + z_buffer_mem = clCreateBuffer(context, CL_MEM_READ_ONLY , sizeof(float) * screen_height * screen_width, NULL, &err); + nfaces_mem = clCreateBuffer(context, CL_MEM_READ_ONLY , sizeof(int) , NULL, &err); + uv_buffer = clCreateBuffer(context, CL_MEM_READ_ONLY , uv_size , NULL, &err); + map_size_buffer = clCreateBuffer(context, CL_MEM_READ_ONLY , sizeof(int) * 2 , NULL, &err); + norms_mem = clCreateBuffer(context, CL_MEM_READ_ONLY , norms_size , NULL, &err); + light_dir_buffer = clCreateBuffer(context, CL_MEM_READ_ONLY , sizeof(float) * 3 , NULL, &err); + diffuse_map_buffer = clCreateBuffer(context, CL_MEM_READ_ONLY , sizeof(char) * 3 * map_size[0] * map_size[1] , NULL, &err); + + + err = clSetKernelArg(fragment_shader_kernel, 0, sizeof(cl_mem), &faces_buffer); + err = clSetKernelArg(fragment_shader_kernel, 1, sizeof(cl_mem), &new_vertices_mem); + err = clSetKernelArg(fragment_shader_kernel, 2, sizeof(cl_mem), &pixel_data_buffer); + err = clSetKernelArg(fragment_shader_kernel, 3, sizeof(cl_mem), &screen_width_mem); + err = clSetKernelArg(fragment_shader_kernel, 4, sizeof(cl_mem), &z_buffer_mem); + err = clSetKernelArg(fragment_shader_kernel, 5, sizeof(cl_mem), &nfaces_mem); + err = clSetKernelArg(fragment_shader_kernel, 6, sizeof(cl_mem), &uv_buffer); + err = clSetKernelArg(fragment_shader_kernel, 7, sizeof(cl_mem), &map_size_buffer); + err = clSetKernelArg(fragment_shader_kernel, 8, sizeof(cl_mem), &light_dir_buffer); + err = clSetKernelArg(fragment_shader_kernel, 9, sizeof(cl_mem), &norms_mem); + err = clSetKernelArg(fragment_shader_kernel, 10, sizeof(cl_mem), &diffuse_map_buffer); + + err = clEnqueueWriteBuffer(commands, screen_width_mem , CL_FALSE, 0, sizeof(int) , &screen_width , 0, NULL, NULL); + err = clEnqueueWriteBuffer(commands, faces_buffer , CL_FALSE, 0, sizeof(int) * 3 * 3 * nfaces , faces , 0, NULL, NULL); + err = clEnqueueWriteBuffer(commands, nfaces_mem , CL_FALSE, 0, sizeof(int) , &nfaces , 0, NULL, NULL); + err = clEnqueueWriteBuffer(commands, uv_buffer , CL_FALSE, 0, uv_size , uv , 0, NULL, NULL); + err = clEnqueueWriteBuffer(commands, map_size_buffer , CL_FALSE, 0, sizeof(int) * 2 , map_size , 0, NULL, NULL); + err = clEnqueueWriteBuffer(commands, light_dir_buffer , CL_FALSE, 0, sizeof(float) * 3 , light_dir , 0, NULL, NULL); + err = clEnqueueWriteBuffer(commands, norms_mem , CL_FALSE, 0, norms_size , norms , 0, NULL, NULL); + err = clEnqueueWriteBuffer(commands, diffuse_map_buffer, CL_FALSE, 0, sizeof(char) * 3 * map_size[0] * map_size[1], diffuse_map , 0, NULL, NULL); + + hidden_pixel_buffer = new int[screen_width*screen_height]; + fragment_shader_buffers_initialized = true; + } + + clear(&pixel_data_buffer, sizeof(int) * screen_width * screen_height, 0); + clear(&z_buffer_mem, sizeof(float) * screen_width * screen_height, 0); + + + size_t fragment_shader_global[] = { nfaces }; + + err = clEnqueueNDRangeKernel(commands, fragment_shader_kernel, 1, NULL, fragment_shader_global, NULL, 0, NULL, NULL); + + err = clEnqueueReadBuffer(commands, pixel_data_buffer, CL_TRUE, 0, sizeof(int) * screen_height * screen_width, hidden_pixel_buffer, 0, NULL, NULL); + + for (int j = 0; j < screen_height; j++) + for (int i = 0; i < screen_width; i++) + set_pixel(i, j, hidden_pixel_buffer[j * screen_width + i]); + + - clReleaseMemObject(mat_z); - clReleaseMemObject(vertices_mem); - clReleaseMemObject(new_vertices_mem); } void destroy_kernels() { clReleaseProgram(vertex_shader_prog); clReleaseKernel(vertex_shader_kernel); -// clReleaseMemObject(mat_z); -// clReleaseMemObject(vertices_mem); -// clReleaseMemObject(new_vertices_mem); + if (vertex_shader_buffers_initialized) { + clReleaseMemObject(mat_z); + clReleaseMemObject(vertices_mem); + clReleaseMemObject(new_vertices_mem); + } + + if (fragment_shader_buffers_initialized) { + clReleaseMemObject(faces_buffer); + clReleaseMemObject(screen_width_mem); + clReleaseMemObject(z_buffer_mem); + clReleaseMemObject(pixel_data_buffer); + clReleaseMemObject(screen_width_mem); + clReleaseMemObject(nfaces_mem); + clReleaseMemObject(uv_buffer); + clReleaseMemObject(map_size_buffer); + clReleaseMemObject(light_dir_buffer); + clReleaseMemObject(norms_mem); + clReleaseMemObject(diffuse_map_buffer); + } clReleaseCommandQueue(commands); - + free(hidden_pixel_buffer); clReleaseContext(context); diff --git a/OpenWindow/kernels.h b/OpenWindow/kernels.h index da5d50d..16855dd 100644 --- a/OpenWindow/kernels.h +++ b/OpenWindow/kernels.h @@ -26,20 +26,29 @@ extern cl_command_queue commands; extern int err; - -#define DIM 2 - -extern size_t mat_mul_global[DIM]; -extern size_t mat_mul_local[DIM]; - extern const char* matrix_mul_kernel_source; // End of: Matrix Multiplication Variables extern const char* vertex_shader_kernel_source; +extern const char* fragment_shader_kernel_source; void init_kernels(); -void mat4_mul(float* A, float* B, float* C); void vertex_shader(float* z, float* vertices, int vertex_count, float* new_vertices); +void fragment_shader( + int* faces, + int nfaces, + float* uv, + size_t uv_size, + float* uniform_m, + float* uniform_mit, + float* light_dir, + unsigned char* diffuse_map, + float* norms, + size_t norms_size, + unsigned char* spec_map, + int* map_size +); + void destroy_kernels(); diff --git a/OpenWindow/main.cpp b/OpenWindow/main.cpp index 8f79185..4a63c41 100644 --- a/OpenWindow/main.cpp +++ b/OpenWindow/main.cpp @@ -78,7 +78,10 @@ void CALLBACK FixedUpdate(HWND hwnd, UINT message, UINT uInt, DWORD dWord) TIME += 0.167; HandleButtonPressed(); camera.ApplyChanges(); + clock_t start = clock(); render(); + clock_t end = clock(); + printf("FPS: %f\n", 1/((float)(end-start)/CLOCKS_PER_SEC)); Update(); } diff --git a/OpenWindow/model.h b/OpenWindow/model.h index 58a12d5..196dd6b 100644 --- a/OpenWindow/model.h +++ b/OpenWindow/model.h @@ -7,11 +7,11 @@ class Model { private: + void load_texture(std::string filename, const char *suffix, TGAImage &img); +public: TGAImage diffusemap_; TGAImage normalmap_; TGAImage specularmap_; - void load_texture(std::string filename, const char *suffix, TGAImage &img); -public: std::vector verts_; std::vector > faces_; // attention, this Vec3i means vertex/uv/normal std::vector norms_; diff --git a/OpenWindow/renderer.cpp b/OpenWindow/renderer.cpp index c59352b..52aa4aa 100644 --- a/OpenWindow/renderer.cpp +++ b/OpenWindow/renderer.cpp @@ -13,10 +13,10 @@ #define VERTICAL_CAMERA_SPEED 0.1 #define VERTICAL_CAMERA_CLAMP_UP 90 #define VERTICAL_CAMERA_CLAMP_DOWN -90 -#define NEAR_CLIP_PLANE 0.3f -#define FAR_CLIP_PLANE 10.0f -#define FOV 30 -#define CAMERA_MOVEMENT_SPEED 0.05f +#define NEAR_CLIP_PLANE 1.f +#define FAR_CLIP_PLANE 200.0f +#define FOV 50 +#define CAMERA_MOVEMENT_SPEED .1f #define DEFAULT_CAMERA_POS Vec3f(0, 0, 5) #define DEFAULT_CAMERA_ROT Vec3f(0, 0, 0) #define LIGHT_INTENSITY 1.5 @@ -36,6 +36,9 @@ Camera camera; Vec3f light_dir = Vec3f(1, 1, 1).normalize(); float* new_verts = (float*)malloc(4 * sizeof(float) * model->nverts()); +int* faces = (int*)malloc(3 * 3 * sizeof(int) * model->nfaces()); + +bool init_flag = false; void init_camera() { camera.SetPosition(DEFAULT_CAMERA_POS); @@ -82,7 +85,7 @@ struct TextureShader : public IShader { TGAColor c = model->diffuse(uv); color = c; for (int i = 0; i < 3; i++) - color[i] = std::fmin(2 + c[i] * (( 1 * diff_intensity + 1 * spec_intensity)), 255) * LIGHT_INTENSITY; + color[i] = std::fmin(3 + c[i] * ((1 * diff_intensity + 0.1 * spec_intensity)), 255);// *LIGHT_INTENSITY; return false; } }; @@ -90,12 +93,17 @@ struct TextureShader : public IShader { void render() { - { + if (!init_flag) { //light_dir = camera.GetForward().normalize() * -1; + viewport(0, 0, screen_width, screen_height, FAR_CLIP_PLANE, NEAR_CLIP_PLANE); + for (int i = 0; i < model->nfaces(); i++) + for (int j = 0; j < 3; j++) + for (int k = 0; k < 3; k++) + faces[i * 9 + j * 3 + k] = model->faces_[i][j][k]; + init_flag = true; } { - viewport(0, 0, screen_width, screen_height, FAR_CLIP_PLANE, NEAR_CLIP_PLANE); Projection = camera.GetProjectionMatrix(); ModelView = camera.GetModelViewMatrix(); } @@ -105,31 +113,79 @@ void render() //model->ApplyTransform(); } - clear_zbuffer(); - TextureShader shader; - shader.uniform_m = (Projection); - shader.uniform_mit = (Projection).invert_transpose(); +// float* normal_test = *(float**)((Vec3f*)&model->norms_); +// +// Vec3f smth = Vec3f( +// normal_test[3 * faces[0 * 9 + 0 * 3 + 2] + 0], +// normal_test[3 * faces[0 * 9 + 0 * 3 + 2] + 1], +// normal_test[3 * faces[0 * 9 + 0 * 3 + 2] + 2] +// ).normalize(); +// printf("Real: %f, %f, %f\n", model->normal(0, 0)[0], model->normal(0, 0)[1], model->normal(0, 0)[2]); +// printf("Please: %f, %f, %f\n", smth[0], smth[1], smth[2]); + + //TextureShader shader; + //shader.uniform_m = (Projection); + //shader.uniform_mit = (Projection).invert_transpose(); Matrix z = ViewPort * Projection * ModelView * model->Transform; + Matrix uniform_m = (Projection); + Matrix uniform_mit = (Projection).invert_transpose(); // Vertex Shader: Should be called per model vertex_shader((float*)&z, *(float**)((Vec3f*) &model->verts_), model->nverts(), new_verts); - #pragma omp parallel for - for (int i = 0; i < model->nfaces(); i++) { - Vec4f screen_coords[3]; - bool out = true; - #pragma omp parallel for - for (int j = 0; j < 3; j++) { - screen_coords[j] = ((Vec4f*)new_verts)[model->faces_[i][j][0]]; - Vec3f screen3(screen_coords[j]); + // Things needed in the GPU fragment shader + // [x] model->faces_ + // [x] model->nfaces() + // [x] new_verts + // [x] model->nverts() + // [x] screen_width + // [x] screen_height + // [x] z_buffer + // [x] uniform_m + // [x] uniform_mit + // [x] light direction + // [ ] diffuse map + // [ ] normal map + // [ ] specular map + // [ ] pixel_data - shader.varying_uv_coords.set_col(j, model->uv(i, j)); - if (screen3.x > 0 && screen3.x < screen_width && screen3.y > 0 && screen3.y < screen_height) out = false; - } - if(!out) - triangle(screen_coords, shader); - } + int map_size[] = {model->diffusemap_.get_width(), model->diffusemap_.get_height()}; + + fragment_shader( + faces, + model->nfaces(), + *(float**)((Vec2f*)&model->uv_), + sizeof(float) * model->uv_.size() * 2, + (float*) &uniform_m, + (float*) &uniform_mit, + (float*) &light_dir, + model->diffusemap_.data, + *(float**)((Vec3f*)&model->norms_), + sizeof(float) * model->norms_.size() * 3, + model->specularmap_.data, + map_size + ); + + // Here starts the CPU fragment shader + //printf("Here starts the loop\n"); + //#pragma omp parallel for + //for (int i = 0; i < model->nfaces(); i++) { + // Vec4f screen_coords[3]; + // bool out = true; + // #pragma omp parallel for + // for (int j = 0; j < 3; j++) { + // screen_coords[j] = ((Vec4f*)new_verts)[model->faces_[i][j][0]]; + // Vec3f screen3(screen_coords[j]); + + // shader.varying_uv_coords.set_col(j, model->uv(i, j)); + // if (screen3.x > 0 && screen3.x < screen_width && screen3.y > 0 && screen3.y < screen_height) out = false; + // } + // if(!out) + // triangle(screen_coords, shader); + //} + + //printf("that's it\n"); } diff --git a/OpenWindow/tgaimage.h b/OpenWindow/tgaimage.h index 63a944b..16ba86f 100644 --- a/OpenWindow/tgaimage.h +++ b/OpenWindow/tgaimage.h @@ -62,14 +62,13 @@ struct TGAColor { class TGAImage { protected: - unsigned char* data; int width; int height; int bytespp; - bool load_rle_data(std::ifstream &in); bool unload_rle_data(std::ofstream &out); public: + unsigned char* data; enum Format { GRAYSCALE=1, RGB=3, RGBA=4 }; diff --git a/OpenWindow/util_renderer.cpp b/OpenWindow/util_renderer.cpp index 49715fa..b789bc5 100644 --- a/OpenWindow/util_renderer.cpp +++ b/OpenWindow/util_renderer.cpp @@ -1,5 +1,6 @@ #include "util_renderer.h" #include "util_window.h" +#include "kernels.h" IShader::~IShader() {} @@ -16,7 +17,7 @@ void viewport(int x, int y, int w, int h, int far_plane, int near_plane) { } int color_to_int(TGAColor col) { - return (col[2] << 16) | (col[1] << 8) | col[0]; + return (col[0] << 24) | (col[1] << 16) | (col[2] << 8) | col[3]; } Vec3f barycentric(Vec3f* pts, Vec3f P) @@ -52,22 +53,38 @@ void triangle( Vec4f* pts, IShader &shader) bounding_box_max[j] = std::fmin(clamp[j], std::fmax(bounding_box_max[j], (int)pts3[i][j])); } + //printf("%f, %f, %f\n", pts3[0][0], pts3[0][1], pts3[0][2]); + //float* points_test = (float*)&pts3; + + //printf("%f, %f, %f\n", points_test[0], points_test[1], points_test[2]); + //printf("%d, %d\n", bounding_box_max[0], bounding_box_max[1]); + + //fragment_shader_pixelwise( + // *(float**)((Vec3f*)&pts3), + // *(int**)((Vec2i*)&bounding_box_min), + // *(int**)((Vec2i*)&bounding_box_max)); + + //fragment_shader_pixelwise( + // (float*)(&pts3), + // (int*)(&bounding_box_min), + // (int*)(&bounding_box_max)); Vec3i P; #pragma omp parallel for for (P.x = bounding_box_min.x; P.x <= bounding_box_max.x; P.x++) { for (P.y = bounding_box_min.y; P.y <= bounding_box_max.y; P.y++) { - Vec3f bc_coord = barycentric(pts3, P); - float frag_depth = 0; - for (int i = 0; i < 3; i++) - frag_depth += pts3[i][2] * bc_coord[i]; - if (bc_coord.x < 0 || bc_coord.y < 0 || bc_coord.z < 0 || z_buffer[ P.x + P.y * screen_width ]>frag_depth) continue; - TGAColor color; - bool discard = shader.fragment(bc_coord, color); - if (!discard) { - z_buffer[P.x + P.y * screen_width] = frag_depth; - set_pixel(P.x, P.y, color_to_int(color)); - } + //Vec3f bc_coord = barycentric(pts3, P); + //float frag_depth = 0; + //for (int i = 0; i < 3; i++) + // frag_depth += pts3[i][2] * bc_coord[i]; + //if (bc_coord.x < 0 || bc_coord.y < 0 || bc_coord.z < 0 || z_buffer[ P.x + P.y * screen_width ]>frag_depth) continue; + //TGAColor color; + //bool discard = shader.fragment(bc_coord, color); + //if (!discard) { + // z_buffer[P.x + P.y * screen_width] = frag_depth; + // set_pixel(P.x, P.y, color_to_int(color)); + set_pixel(P.x, P.y, 0x00ff00); + //} } } } \ No newline at end of file diff --git a/OpenWindow/util_window.cpp b/OpenWindow/util_window.cpp index 7509429..7e6693f 100644 --- a/OpenWindow/util_window.cpp +++ b/OpenWindow/util_window.cpp @@ -95,9 +95,9 @@ void destroy_window() { void set_pixel(unsigned int x, unsigned int y, unsigned int color) { unsigned long pixel_index = y * bytes_per_row + x * 3; - pixel_data[pixel_index + 0] = (char)(color >> 0); + pixel_data[pixel_index + 2] = (char)(color >> 0); pixel_data[pixel_index + 1] = (char)(color >> 8); - pixel_data[pixel_index + 2] = (char)(color >> 16); + pixel_data[pixel_index + 0] = (char)(color >> 16); if (!screen_changed) screen_changed = true;