Skip to content
New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Update wgpu version #11

Merged
merged 2 commits into from
Jan 1, 2024
Merged
Changes from all commits
Commits
File filter

Filter by extension

Filter by extension


Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
1,396 changes: 863 additions & 533 deletions Cargo.lock

Large diffs are not rendered by default.

4 changes: 2 additions & 2 deletions Cargo.toml
Original file line number Diff line number Diff line change
@@ -11,8 +11,8 @@ log = "0.4"
noise = "0.8.2"
pollster = "0.3.0"
regex = "1.8.4"
wgpu = "0.16.0"
winit = "0.28.6"
wgpu = "0.18.0"
winit = { version = "0.29.7", features = ["rwh_05"] }

[dev-dependencies]
futures-intrusive = "0.5.0"
132 changes: 21 additions & 111 deletions src/app.rs
Original file line number Diff line number Diff line change
@@ -1,55 +1,31 @@
use winit::{
event::*,
event_loop::{ControlFlow, EventLoop},
window::{Window, WindowBuilder},
};

use crate::{heat_equation::HeatEquation, renderer::Renderer};
use winit::window::Window;

struct App {
pub struct App {
surface: wgpu::Surface,
device: wgpu::Device,
queue: wgpu::Queue,
config: wgpu::SurfaceConfiguration,
size: winit::dpi::PhysicalSize<u32>,
window: Window,
heat_eqn: HeatEquation,
renderer: Renderer,
}

impl App {
// Creating some of the wgpu types requires async code
async fn new(window: Window) -> Self {
pub async fn new(window: &Window) -> Self {
let size = window.inner_size();
// The instance is a handle to our GPU
// Backends::all => Vulkan + Metal + DX12 + Browser WebGPU
let instance = wgpu::Instance::new(wgpu::InstanceDescriptor {
backends: wgpu::Backends::all(),
dx12_shader_compiler: Default::default(),
..Default::default()
});

// # Safety
//
// The surface needs to live as long as the window that created it.
// State owns the window so this should be safe.
let (size, surface) = unsafe {
let size = window.inner_size();

#[cfg(any(not(target_arch = "wasm32"), target_os = "emscripten"))]
let surface = instance.create_surface(&window).unwrap();
#[cfg(all(target_arch = "wasm32", not(target_os = "emscripten")))]
let surface = {
if let Some(offscreen_canvas_setup) = &offscreen_canvas_setup {
log::info!("Creating surface from OffscreenCanvas");
instance.create_surface_from_offscreen_canvas(
offscreen_canvas_setup.offscreen_canvas.clone(),
)
} else {
instance.create_surface(&window)
}
}
.unwrap();

(size, surface)
let surface = unsafe {
instance
.create_surface(&window)
.expect("Surface creation failed")
};

let adapter = instance
@@ -133,7 +109,6 @@ impl App {
let renderer = Renderer::new(&device, &config, texture_view);

Self {
window,
surface,
device,
queue,
@@ -144,11 +119,7 @@ impl App {
}
}

pub fn window(&self) -> &Window {
&self.window
}

fn resize(&mut self, new_size: winit::dpi::PhysicalSize<u32>) {
pub fn resize(&mut self, new_size: winit::dpi::PhysicalSize<u32>) {
if new_size.width > 0 && new_size.height > 0 {
self.size = new_size;
self.config.width = new_size.width;
@@ -157,85 +128,24 @@ impl App {
}
}

fn compute_step(&mut self) {
/// Specific method created for use
/// when there's a `WindowEvent::RedrawRequested` and
/// a `Err(wgpu::SurfaceError::Lost)` occurs.
pub fn reacquire_size(&mut self) {
let new_size = self.size;
self.resize(new_size);
}

pub fn update(&mut self) {
self.heat_eqn.compute_step(&self.device, &self.queue);
}

fn render(&self) -> Result<(), wgpu::SurfaceError> {
pub fn render(&self, window: &Window) -> Result<(), wgpu::SurfaceError> {
self.renderer
.render(&self.device, &self.surface, &self.queue)
.render(window, &self.device, &self.surface, &self.queue)
}
}

pub async fn run() {
env_logger::init();
let event_loop = EventLoop::new();
let window = WindowBuilder::new().build(&event_loop).unwrap();
// let mut app = App::new(window).await;
// create app to be used in two threads
let mut app = App::new(window).await;

event_loop.run(move |event, _, control_flow| {
match event {
Event::RedrawEventsCleared => {
// TODO: in the wgpu example, an async executor is used for polling
// we need to check how to sync the compute step with the render step
// spawner.run_until_stalled();

app.window().request_redraw();
}
Event::WindowEvent {
event:
WindowEvent::Resized(size)
| WindowEvent::ScaleFactorChanged {
new_inner_size: &mut size,
..
},
..
} => {
log::info!("Resizing to {:?}", size);
app.resize(size);
app.surface.configure(&app.device, &app.config);
}
Event::WindowEvent { event, .. } => match event {
WindowEvent::CloseRequested => {
*control_flow = ControlFlow::Exit;
}
WindowEvent::KeyboardInput {
input:
winit::event::KeyboardInput {
virtual_keycode: Some(keycode),
state: ElementState::Pressed,
..
},
..
} => {
if keycode == VirtualKeyCode::Escape {
*control_flow = ControlFlow::Exit
}
}
_ => {
// TODO example.update(event);
}
},
Event::RedrawRequested(window_id) if window_id == app.window().id() => {
app.compute_step();

match app.render() {
Ok(_) => {}
// Reconfigure the surface if lost
Err(wgpu::SurfaceError::Lost) => app.resize(app.size),
// The system is out of memory, we should probably quit
Err(wgpu::SurfaceError::OutOfMemory) => *control_flow = ControlFlow::Exit,
// All other errors (Outdated, Timeout) should be resolved by the next frame
Err(e) => eprintln!("{:?}", e),
}
}
_ => {}
}
});
}

fn gaussian(x: f32) -> f32 {
const SIGMA: f32 = 0.5;
const SIGMA2: f32 = SIGMA * SIGMA;
1 change: 1 addition & 0 deletions src/compute.rs
Original file line number Diff line number Diff line change
@@ -115,6 +115,7 @@ impl Compute {
);
let mut compute_pass = encoder.begin_compute_pass(&wgpu::ComputePassDescriptor {
label: Some("Heat pass"),
timestamp_writes: None,
});
compute_pass.set_pipeline(&self.pipeline);
compute_pass.set_bind_group(0, self.bind_group.get(direction), &[]);
11 changes: 8 additions & 3 deletions src/conjugate_gradient.rs
Original file line number Diff line number Diff line change
@@ -108,7 +108,10 @@ impl CG {
label: Some("Conjugate Gradient"),
});
// Initialize r = b - A * x
let cdescriptor = wgpu::ComputePassDescriptor { label: None };
let cdescriptor = wgpu::ComputePassDescriptor {
label: None,
timestamp_writes: None,
};
let mut compute_pass = encoder.begin_compute_pass(&cdescriptor);
for stage in self.init_stages.iter() {
stage.add_to_pass(&mut compute_pass);
@@ -117,8 +120,10 @@ impl CG {
encoder.copy_buffer_to_buffer(r, 0, p, 0, r.size());
// describes all the stages in a single iteration of the CG algorithm
for _ in 0..self.max_steps {
let mut compute_pass =
encoder.begin_compute_pass(&wgpu::ComputePassDescriptor { label: None });
let mut compute_pass = encoder.begin_compute_pass(&wgpu::ComputePassDescriptor {
label: None,
timestamp_writes: None,
});
for s in self.stages.iter() {
s.add_to_pass(&mut compute_pass);
}
2 changes: 2 additions & 0 deletions src/heat_equation.rs
Original file line number Diff line number Diff line change
@@ -167,6 +167,7 @@ impl HeatEquation {
});
let mut compute_pass = encoder.begin_compute_pass(&wgpu::ComputePassDescriptor {
label: Some("Initial SpMV Compute Pass (tmp = B*U)"),
timestamp_writes: None,
});
if self.iteration % 2 == 0 {
self.initial_spmv_forward.add_to_pass(&mut compute_pass);
@@ -190,6 +191,7 @@ impl HeatEquation {
});
let mut compute_pass = encoder.begin_compute_pass(&wgpu::ComputePassDescriptor {
label: Some("Write to Texture Compute Pass"),
timestamp_writes: None,
});
if self.iteration % 2 == 0 {
self.write_to_texture_forward.add_to_pass(&mut compute_pass);
59 changes: 57 additions & 2 deletions src/main.rs
Original file line number Diff line number Diff line change
@@ -1,5 +1,60 @@
use heat_wgpu::app::run;
use heat_wgpu::app::App;
use winit::{
event::{Event, WindowEvent},
event_loop::EventLoop,
keyboard::{KeyCode, PhysicalKey::Code},
window::WindowBuilder,
};

fn main() {
pollster::block_on(run());
env_logger::init();
let event_loop = EventLoop::new().expect("Event loop creation failed");
let window = WindowBuilder::new()
.build(&event_loop)
.expect("Window builder creation failed");
let mut app = pollster::block_on(App::new(&window));

event_loop
.run(move |event, event_loop_window_target| {
match event {
Event::WindowEvent {
ref event,
window_id,
} if window_id == window.id() => {
match &event {
WindowEvent::Resized(physical_size) => app.resize(*physical_size),
WindowEvent::ScaleFactorChanged { .. } => {
app.resize(window.inner_size());
}
WindowEvent::CloseRequested => event_loop_window_target.exit(),
WindowEvent::KeyboardInput { event, .. } => {
if event.state.is_pressed()
&& matches!(event.physical_key, Code(KeyCode::Escape))
{
event_loop_window_target.exit();
}
}
WindowEvent::RedrawRequested => {
app.update();

match app.render(&window) {
Ok(_) => {}
// Recreate the swap_chain if lost
Err(wgpu::SurfaceError::Lost) => app.reacquire_size(),
// The system is out of memory, we should probably quit
Err(wgpu::SurfaceError::OutOfMemory) => {
event_loop_window_target.exit()
}
// All other errors (Outdated, Timeout) should be resolved by the next frame
Err(e) => eprintln!("Unhandled error: {:?}", e),
}
}
_ => {}
}
}
Event::AboutToWait => window.request_redraw(),
_ => {}
}
})
.unwrap();
}
8 changes: 7 additions & 1 deletion src/renderer.rs
Original file line number Diff line number Diff line change
@@ -1,5 +1,6 @@
use crate::vertex::Vertex;
use wgpu::util::DeviceExt;
use winit::window::Window;

const VERTICES: &[Vertex] = &[
crate::vertex!([-1.0, 1.0, 0.0], [0.0, 0.0]), // top left
@@ -144,6 +145,7 @@ impl Renderer {
}
pub fn render(
&self,
window: &Window,
device: &wgpu::Device,
surface: &wgpu::Surface,
queue: &wgpu::Queue,
@@ -162,10 +164,12 @@ impl Renderer {
resolve_target: None,
ops: wgpu::Operations {
load: wgpu::LoadOp::Clear(wgpu::Color::BLACK),
store: true,
store: wgpu::StoreOp::Store,
},
})],
depth_stencil_attachment: None,
timestamp_writes: None,
occlusion_query_set: None,
});
render_pass.set_pipeline(&self.pipeline);
render_pass.set_bind_group(0, &self.bind_group, &[]);
@@ -174,6 +178,8 @@ impl Renderer {
render_pass.draw_indexed(0..self.num_indices, 0, 0..1);
drop(render_pass);

window.pre_present_notify();

queue.submit(std::iter::once(encoder.finish()));
output.present();

6 changes: 4 additions & 2 deletions src/shader_tests/spmv.rs
Original file line number Diff line number Diff line change
@@ -162,8 +162,10 @@ mod tests {
let mut encoder =
device.create_command_encoder(&wgpu::CommandEncoderDescriptor { label: None });
{
let mut cpass =
encoder.begin_compute_pass(&wgpu::ComputePassDescriptor { label: None });
let mut cpass = encoder.begin_compute_pass(&wgpu::ComputePassDescriptor {
label: None,
timestamp_writes: None,
});
cpass.set_pipeline(&compute_pipeline);
cpass.set_bind_group(0, &bind_group, &[]);
cpass.insert_debug_marker("compute sparse matrix-vector multiplication (SpMV)");
6 changes: 4 additions & 2 deletions src/shader_tests/sum_reduce.rs
Original file line number Diff line number Diff line change
@@ -179,8 +179,10 @@ mod tests {
let mut encoder =
device.create_command_encoder(&wgpu::CommandEncoderDescriptor { label: None });
{
let mut cpass =
encoder.begin_compute_pass(&wgpu::ComputePassDescriptor { label: None });
let mut cpass = encoder.begin_compute_pass(&wgpu::ComputePassDescriptor {
label: None,
timestamp_writes: None,
});
cpass.set_pipeline(&compute_pipeline_1);
cpass.set_bind_group(0, &bind_group_1, &[]);
cpass.insert_debug_marker("compute vector block sum");
6 changes: 4 additions & 2 deletions src/shader_tests/vec_mul.rs
Original file line number Diff line number Diff line change
@@ -131,8 +131,10 @@ mod tests {
let mut encoder =
device.create_command_encoder(&wgpu::CommandEncoderDescriptor { label: None });
{
let mut cpass =
encoder.begin_compute_pass(&wgpu::ComputePassDescriptor { label: None });
let mut cpass = encoder.begin_compute_pass(&wgpu::ComputePassDescriptor {
label: None,
timestamp_writes: None,
});
cpass.set_pipeline(&compute_pipeline);
cpass.set_bind_group(0, &bind_group, &[]);
cpass.insert_debug_marker("compute vector element-wise multiplication");
2 changes: 1 addition & 1 deletion src/shaders/dot.wgsl
Original file line number Diff line number Diff line change
@@ -1,6 +1,6 @@
@group(0) @binding(0) var input_vec_a: texture_storage_2d<r32float, read>;
@group(0) @binding(1) var input_vec_b: texture_storage_2d<r32float, read>;
@group(0) @binding(2) var<storage, write> output: f32;
@group(0) @binding(2) var<storage, read_write> output: f32;

@compute @workgroup_size(16)
fn main(@builtin(global_invocation_id) global_id: vec3<u32>) {
Loading