moving everything to global memory to support host side animations

This commit is contained in:
Brett Weiland 2024-12-24 20:51:41 -06:00
parent c05aa71f40
commit 04465fde84
7 changed files with 49 additions and 34 deletions

View File

@ -1,4 +1,5 @@
#include <limits>
#include "scene.cuh"
#include "camera.cuh"
__device__ void Camera::render(Scene scene, uint8_t *image) {
@ -11,7 +12,8 @@ __device__ void Camera::render(Scene scene, uint8_t *image) {
Render_object *closest_obj;
T total_dist = 0;
Color pcolor;
for(int steps = 0; steps < max_steps; steps++) {
for(int steps = 0; steps < max_steps; steps++) { //this could be heavily optimized.
scene.raycast(ray);
T min_dist = T_MAX;
T dist;
ray.start = this->pos_ + (total_dist * ray.direction);

View File

@ -1,5 +1,6 @@
#ifndef CAMERA_H
#define CAMERA_H
#include <stdint.h>
#include <limits>
#include "entity.cuh"
@ -9,10 +10,9 @@
//template <class T> class scene;
//I am soooo high lol
class Camera : public Entity {
public:
__device__ Camera(const T fov = 1, const vect3 pos = make_vect3(0), const vect3 rot = make_vect3(0))
Camera(const T fov = 1, const vect3 pos = make_vect3(0), const vect3 rot = make_vect3(0))
: fov(fov), Entity(pos, rot, make_vect3(0)) {};
__device__ void render(Scene scene, uint8_t *image);
private:
@ -23,6 +23,7 @@ class Camera : public Entity {
float2 size;
};
/**
//later we'll make scenes objects, rn im lazy (TODO)
template <class T> __device__ void camera::render() {

View File

@ -8,9 +8,9 @@
//out of the question
class Entity {
public:
__device__ Entity() : pos_(make_vect3(0)), rot_(make_vect3(0)), scale_(make_vect3(0)) {};
__device__ Entity(const vect3 pos, const vect3 rot, const vect3 scale) : pos_(pos), rot_(rot), scale_(scale) {};
__device__ Entity(const float3 pos) : pos_(pos), rot_(make_vect3(0)), scale_(make_vect3(0)) {};
Entity() : pos_(make_vect3(0)), rot_(make_vect3(0)), scale_(make_vect3(0)) {};
Entity(const vect3 pos, const vect3 rot, const vect3 scale) : pos_(pos), rot_(rot), scale_(scale) {};
Entity(const float3 pos) : pos_(pos), rot_(make_vect3(0)), scale_(make_vect3(0)) {};
vect3 get_pos() const { return pos_; }

View File

@ -5,8 +5,8 @@
#include "camera.cuh"
__global__ void render(uint8_t *image) {
Scene scene;
Camera cam(1, make_vect3(0,0,0));
cam.render(scene, image);
//Scene scene;
//Camera cam(1, make_vect3(0,0,0));
//cam.render(scene, image);
//scene.render(image);
}

View File

@ -3,9 +3,18 @@
#include <stdint.h>
#include <cuda_runtime.h>
#include <string.h>
#include <raylib.h>
namespace raylib {
#include <raylib.h>
}
#include "kernel.cuh"
#include "scene.cuh"
#include "sphere.cuh"
#include "camera.cuh"
int main() {
//bluuuuugh i'll figure out occupancy later, this res are easy
@ -17,7 +26,7 @@ int main() {
uint8_t *image_d;
Color texture_data[res_x * res_y];
SetTargetFPS(10);
raylib::SetTargetFPS(60);
//see if GPU is connected (my egpu is finicky)
@ -30,14 +39,20 @@ int main() {
}
}
SetTraceLogLevel(LOG_ERROR);
InitWindow(res_x, res_y, "cuda teseteroni");
raylib::SetTraceLogLevel(raylib::LOG_ERROR);
raylib::InitWindow(res_x, res_y, "cuda teseteroni");
//TODO could probably cut out
Image image = GenImageColor(res_x, res_y, BLUE);
Texture tex = LoadTextureFromImage(image);
raylib::Image image = raylib::GenImageColor(res_x, res_y, raylib::BLUE);
raylib::Texture tex = raylib::LoadTextureFromImage(image);
while(!WindowShouldClose()) {
//initilize scene
Scene scene;
Sphere sp1 = Sphere(make_vect3(0, -2, 5));
Sphere sp2 = Sphere(make_vect3(2, 0, 5));
Camera cam = Camera(1, make_float3(0,0,0));
while(!raylib::WindowShouldClose()) {
cudaError_t err;
//cuda stuff
cudaMalloc((void **)&image_d, res_x * res_y * sizeof(Color));
@ -48,11 +63,12 @@ int main() {
cudaDeviceSynchronize();
cudaMemcpy(texture_data, (void **)image_d, res_x * res_y * sizeof(Color), cudaMemcpyDeviceToHost);
BeginDrawing();
UpdateTexture(tex, texture_data);
DrawTexture(tex, 0, 0, WHITE);
DrawFPS(0, 0);
EndDrawing();
raylib::BeginDrawing();
raylib::ClearBackground(raylib::BLACK);
raylib::UpdateTexture(tex, texture_data);
raylib::DrawTexture(tex, 0, 0, raylib::WHITE);
raylib::DrawFPS(0, 0);
raylib::EndDrawing();
}
return 0;
}

View File

@ -1,15 +1,9 @@
#include <stdio.h>
#include "scene.cuh"
__device__ void Scene::debug() {
//const uint3 unnormalized_uv = ((blockDim * blockIdx) + threadIdx);
//const unsigned int img_index = (unnormalized_uv.x + (unnormalized_uv.y * blockDim.x * gridDim.x)) * 4;
}
__device__ Color Scene::raycast(struct Ray ray) {
if(ray.bounces++ > max_bounces) return make_color(0);
__device__ Render_object *Scene::raycast(struct Ray ray) {
if(ray.bounces++ > max_bounces);
const size_t obj_count = sizeof(objs) / sizeof(objs[0]);
for(size_t obj_i = 0; obj_i < obj_count; obj_i++);
return make_color(0);
}

View File

@ -5,24 +5,26 @@
#include "render_object.cuh"
#include "sphere.cuh"
#include "include/helper_math.h"
//#include "camera.cuh"
#include <stdint.h>
//template <class T> class camera;
class Camera;
//when we get animations with multiple scenes, we'll make this a virtual function
//with array of DE objects and cam
class Scene {
const int max_bounces = 10;
public:
__device__ void debug();
__device__ Color raycast(struct Ray ray);
__device__ Render_object *raycast(struct Ray ray);
__device__ Render_object **get_objects() { return objs; }
__device__ Camera *get_camera() { return active_cam; }
private:
Sphere sp1 = Sphere(make_vect3(0, -2, 5));
Sphere sp2 = Sphere(make_vect3(2, 0, 5));
Sphere sp1 = Sphere(make_vect3(0, -2, 5)); //TODO move to it's own file or whatever
Sphere sp2 = Sphere(make_vect3(2, 0, 5)); //do some like "add_object" function or somethin
protected:
//idk why I need to specify the size... why can't the compiler figure that out?
Render_object *objs[3] = {&sp1, &sp2, nullptr};
Camera *active_cam;
uint8_t *image;
};