diff --git a/examples-android/Cargo.toml b/examples-android/Cargo.toml index 5dcff8f0..8c05e4a5 100644 --- a/examples-android/Cargo.toml +++ b/examples-android/Cargo.toml @@ -21,6 +21,7 @@ ndk-glue = "0.7" ndk-context = "0.1" android-activity = { version = "0.6", features = ["native-activity"] } android_logger = "0.14" +del-msh-core = "=0.1.33" [lib] name = "xr" diff --git a/examples-android/xr.rs b/examples-android/xr.rs index d6ff8378..80a88173 100644 --- a/examples-android/xr.rs +++ b/examples-android/xr.rs @@ -1,6 +1,8 @@ #![cfg(target_os = "android")] +#![allow(irrefutable_let_patterns)] use std::{ + mem, ptr, sync::{ atomic::{AtomicBool, Ordering}, Arc, @@ -10,158 +12,239 @@ use std::{ }; use blade_graphics as gpu; -use bytemuck::{Pod, Zeroable}; -use glam::{Mat4, Quat, Vec3}; use log::info; use openxr as xr; const VIEW_TYPE: xr::ViewConfigurationType = xr::ViewConfigurationType::PRIMARY_STEREO; -const MAX_XR_EYES: usize = 2; -const NEAR_Z: f32 = 0.05; +const TORUS_RADIUS: f32 = 1.5; const FAR_Z: f32 = 100.0; +const TARGET_FORMAT: gpu::TextureFormat = gpu::TextureFormat::Rgba8Unorm; +const XR_WORLD_Z_OFFSET: f32 = 8.0; #[repr(C)] -#[derive(Clone, Copy, Pod, Zeroable)] -struct Uniforms { - mvp: [[f32; 4]; 4], - tint: [f32; 4], +#[derive(Clone, Copy, bytemuck::Zeroable, bytemuck::Pod)] +struct Parameters { + cam_position: [f32; 3], + depth: f32, + cam_orientation: [f32; 4], + fov: [f32; 4], // left, right, down, up + torus_radius: f32, + rotation_angle: f32, + pad: [f32; 2], } #[derive(blade_macros::ShaderData)] -struct Params { - globals: gpu::BufferPiece, +struct TraceData { + parameters: Parameters, + acc_struct: gpu::AccelerationStructure, + output: gpu::TextureView, } -fn model_matrix() -> Mat4 { - Mat4::from_translation(Vec3::new(0.0, 0.0, -2.0)) -} - -fn view_matrix_from_pose(pose: gpu::XrPose) -> Mat4 { - let orientation = Quat::from_xyzw( - pose.orientation[0], - pose.orientation[1], - pose.orientation[2], - pose.orientation[3], - ); - let position = Vec3::from_array(pose.position); - Mat4::from_rotation_translation(orientation, position).inverse() -} - -fn projection_matrix_from_fov(fov: gpu::XrFov, near_z: f32, far_z: f32) -> Mat4 { - let tan_left = fov.angle_left.tan(); - let tan_right = fov.angle_right.tan(); - let tan_down = fov.angle_down.tan(); - let tan_up = fov.angle_up.tan(); - - let tan_width = tan_right - tan_left; - let tan_height = tan_up - tan_down; - let z_range = far_z - near_z; - - Mat4::from_cols_array_2d(&[ - [2.0 / tan_width, 0.0, 0.0, 0.0], - [0.0, 2.0 / tan_height, 0.0, 0.0], - [ - (tan_right + tan_left) / tan_width, - (tan_up + tan_down) / tan_height, - -far_z / z_range, - -1.0, - ], - [0.0, 0.0, -(far_z * near_z) / z_range, 0.0], - ]) +#[derive(blade_macros::ShaderData)] +struct DrawData { + input: gpu::TextureView, } struct Example { xr_surface: gpu::XrSurface, - pipeline: gpu::RenderPipeline, + target: gpu::Texture, + target_view: gpu::TextureView, + blas: gpu::AccelerationStructure, + tlas: gpu::AccelerationStructure, + rt_pipeline: gpu::ComputePipeline, + draw_pipeline: gpu::RenderPipeline, command_encoder: gpu::CommandEncoder, - params_buf: gpu::Buffer, - depth_texture: gpu::Texture, - depth_views: [gpu::TextureView; MAX_XR_EYES], + prev_sync_point: Option, start_time: Instant, - rendered_frames: u64, xr_debug: bool, + rendered_frames: u64, + anchor_head_pos: Option<[f32; 3]>, } impl Example { fn new(context: &gpu::Context, xr_debug: bool) -> Self { + let capabilities = context.capabilities(); + assert!(capabilities + .ray_query + .contains(gpu::ShaderVisibility::COMPUTE)); + let xr_surface = context .create_xr_surface() .expect("Unable to create XR surface"); - let color_format = xr_surface.format(); + let extent = xr_surface.extent(); + + let target = context.create_texture(gpu::TextureDesc { + name: "ray-query-target", + format: TARGET_FORMAT, + size: extent, + dimension: gpu::TextureDimension::D2, + array_layer_count: 1, + mip_level_count: 1, + sample_count: 1, + usage: gpu::TextureUsage::RESOURCE | gpu::TextureUsage::STORAGE, + external: None, + }); + let target_view = context.create_texture_view( + target, + gpu::TextureViewDesc { + name: "ray-query-target", + format: TARGET_FORMAT, + dimension: gpu::ViewDimension::D2, + subresources: &gpu::TextureSubresources::default(), + }, + ); let shader = context.create_shader(gpu::ShaderDesc { source: include_str!("xr.wgsl"), }); - let data_layout = ::layout(); - let pipeline = context.create_render_pipeline(gpu::RenderPipelineDesc { - name: "xr", - data_layouts: &[&data_layout], - vertex: shader.at("vs_main"), - vertex_fetches: &[], + shader.check_struct_size::(); + + let rt_layout = ::layout(); + let draw_layout = ::layout(); + let rt_pipeline = context.create_compute_pipeline(gpu::ComputePipelineDesc { + name: "ray-trace", + data_layouts: &[&rt_layout], + compute: shader.at("main"), + }); + let draw_pipeline = context.create_render_pipeline(gpu::RenderPipelineDesc { + name: "ray-query-draw", + data_layouts: &[&draw_layout], primitive: gpu::PrimitiveState { + topology: gpu::PrimitiveTopology::TriangleStrip, ..Default::default() }, - depth_stencil: Some(gpu::DepthStencilState { - format: gpu::TextureFormat::Depth32Float, - depth_write_enabled: true, - depth_compare: gpu::CompareFunction::Less, - stencil: Default::default(), - bias: Default::default(), - }), - fragment: Some(shader.at("fs_main")), - color_targets: &[gpu::ColorTargetState::from(color_format)], - multisample_state: gpu::MultisampleState::default(), + vertex: shader.at("draw_vs"), + vertex_fetches: &[], + fragment: Some(shader.at("draw_fs")), + color_targets: &[gpu::ColorTargetState::from(xr_surface.format())], + depth_stencil: None, + multisample_state: Default::default(), }); - let command_encoder = context.create_command_encoder(gpu::CommandEncoderDesc { - name: "xr", - buffer_count: 1, + + let (indices, vertex_values) = + del_msh_core::trimesh3_primitive::torus_yup::(TORUS_RADIUS, 1.0, 100, 20); + let vertex_buf = context.create_buffer(gpu::BufferDesc { + name: "vertices", + size: (vertex_values.len() * mem::size_of::()) as u64, + memory: gpu::Memory::Shared, }); - let params_buf = context.create_buffer(gpu::BufferDesc { - name: "xr-params", - size: (std::mem::size_of::() * MAX_XR_EYES) as u64, + unsafe { + ptr::copy_nonoverlapping( + vertex_values.as_ptr(), + vertex_buf.data() as *mut f32, + vertex_values.len(), + ) + }; + + let index_buf = context.create_buffer(gpu::BufferDesc { + name: "indices", + size: (indices.len() * mem::size_of::()) as u64, memory: gpu::Memory::Shared, }); - let extent = xr_surface.extent(); - let view_count = xr_surface.view_count() as usize; - let depth_texture = context.create_texture(gpu::TextureDesc { - name: "xr-depth", - format: gpu::TextureFormat::Depth32Float, - size: extent, - dimension: gpu::TextureDimension::D2, - array_layer_count: view_count as u32, - mip_level_count: 1, - usage: gpu::TextureUsage::TARGET, - sample_count: 1, - external: None, + unsafe { + ptr::copy_nonoverlapping(indices.as_ptr(), index_buf.data() as *mut u16, indices.len()) + }; + + let meshes = [gpu::AccelerationStructureMesh { + vertex_data: vertex_buf.at(0), + vertex_format: gpu::VertexFormat::F32Vec3, + vertex_stride: mem::size_of::() as u32 * 3, + vertex_count: vertex_values.len() as u32 / 3, + index_data: index_buf.at(0), + index_type: Some(gpu::IndexType::U16), + triangle_count: indices.len() as u32 / 3, + transform_data: gpu::Buffer::default().at(0), + is_opaque: true, + }]; + + let blas_sizes = context.get_bottom_level_acceleration_structure_sizes(&meshes); + let blas = context.create_acceleration_structure(gpu::AccelerationStructureDesc { + name: "torus-blas", + ty: gpu::AccelerationStructureType::BottomLevel, + size: blas_sizes.data, }); - let mut depth_views = [gpu::TextureView::default(); MAX_XR_EYES]; - for eye in 0..view_count.min(MAX_XR_EYES) { - depth_views[eye] = context.create_texture_view( - depth_texture, - gpu::TextureViewDesc { - name: "xr-depth-eye", - format: gpu::TextureFormat::Depth32Float, - dimension: gpu::ViewDimension::D2, - subresources: &gpu::TextureSubresources { - base_mip_level: 0, - mip_level_count: std::num::NonZeroU32::new(1), - base_array_layer: eye as u32, - array_layer_count: std::num::NonZeroU32::new(1), - }, - }, + + let x_angle = 0.5f32; + let instances = [ + gpu::AccelerationStructureInstance { + acceleration_structure_index: 0, + transform: [ + [1.0, 0.0, 0.0, -1.5], + [0.0, x_angle.cos(), x_angle.sin(), 0.0], + [0.0, -x_angle.sin(), x_angle.cos(), 0.0], + ] + .into(), + mask: 0xFF, + custom_index: 0, + }, + gpu::AccelerationStructureInstance { + acceleration_structure_index: 0, + transform: [ + [1.0, 0.0, 0.0, 1.5], + [0.0, x_angle.sin(), -x_angle.cos(), 0.0], + [0.0, x_angle.cos(), x_angle.sin(), 0.0], + ] + .into(), + mask: 0xFF, + custom_index: 0, + }, + ]; + let tlas_sizes = context.get_top_level_acceleration_structure_sizes(instances.len() as u32); + let instance_buffer = + context.create_acceleration_structure_instance_buffer(&instances, &[blas]); + let tlas = context.create_acceleration_structure(gpu::AccelerationStructureDesc { + name: "torus-tlas", + ty: gpu::AccelerationStructureType::TopLevel, + size: tlas_sizes.data, + }); + let tlas_scratch_offset = + (blas_sizes.scratch | (gpu::limits::ACCELERATION_STRUCTURE_SCRATCH_ALIGNMENT - 1)) + 1; + let scratch_buffer = context.create_buffer(gpu::BufferDesc { + name: "scratch", + size: tlas_scratch_offset + tlas_sizes.scratch, + memory: gpu::Memory::Device, + }); + + let mut command_encoder = context.create_command_encoder(gpu::CommandEncoderDesc { + name: "xr-ray-query", + buffer_count: 1, + }); + command_encoder.start(); + command_encoder.init_texture(target); + if let mut pass = command_encoder.acceleration_structure("build BLAS") { + pass.build_bottom_level(blas, &meshes, scratch_buffer.at(0)); + } + if let mut pass = command_encoder.acceleration_structure("build TLAS") { + pass.build_top_level( + tlas, + &[blas], + instances.len() as u32, + instance_buffer.at(0), + scratch_buffer.at(tlas_scratch_offset), ); } + let sync_point = context.submit(&mut command_encoder); + assert!(context.wait_for(&sync_point, !0)); + + context.destroy_buffer(vertex_buf); + context.destroy_buffer(index_buf); + context.destroy_buffer(instance_buffer); + context.destroy_buffer(scratch_buffer); Self { xr_surface, - pipeline, + target, + target_view, + blas, + tlas, + rt_pipeline, + draw_pipeline, command_encoder, - params_buf, - depth_texture, - depth_views, + prev_sync_point: None, start_time: Instant::now(), - rendered_frames: 0, xr_debug, + rendered_frames: 0, + anchor_head_pos: None, } } @@ -169,64 +252,93 @@ impl Example { let frame = if let Some(frame) = self.xr_surface.acquire_frame(context) { frame } else { - if self.xr_debug { - info!("XR frame: should_render=false (ended without layers)"); - } return; }; self.command_encoder.start(); self.command_encoder.init_texture(frame.texture()); + self.command_encoder.init_texture(self.target); + + let target_extent = self.xr_surface.extent(); + let view_count = frame.xr_view_count(); + let rotation_angle = self.start_time.elapsed().as_secs_f32() * 0.4; + + // Anchor world-space to the initial headset position for stable tracking. + if self.anchor_head_pos.is_none() && view_count >= 2 { + let left = frame.xr_view(0).pose.position; + let right = frame.xr_view(1).pose.position; + self.anchor_head_pos = Some([ + 0.5 * (left[0] + right[0]), + 0.5 * (left[1] + right[1]), + 0.5 * (left[2] + right[2]), + ]); + } + let anchor = self.anchor_head_pos.unwrap_or([0.0; 3]); - let angle = self.start_time.elapsed().as_secs_f32(); - let model = model_matrix() * Mat4::from_rotation_y(angle); - let view_count = frame.xr_view_count().min(MAX_XR_EYES as u32); for eye in 0..view_count { - let xr_view = frame.xr_view(eye); - let view = view_matrix_from_pose(xr_view.pose); - let proj = projection_matrix_from_fov(xr_view.fov, NEAR_Z, FAR_Z); - let mvp = proj * view * model; - let uniforms = Uniforms { - mvp: mvp.to_cols_array_2d(), - tint: [1.0, 1.0, 1.0, 1.0], - }; - let params_offset = (eye as usize * std::mem::size_of::()) as isize; - unsafe { - *(self.params_buf.data().offset(params_offset) as *mut Uniforms) = uniforms; + let view = frame.xr_view(eye); + let cam_position = [ + view.pose.position[0] - anchor[0], + view.pose.position[1] - anchor[1], + view.pose.position[2] - anchor[2] + XR_WORLD_Z_OFFSET, + ]; + if let mut pass = self.command_encoder.compute("ray-trace") { + let groups = self.rt_pipeline.get_dispatch_for(target_extent); + if let mut pc = pass.with(&self.rt_pipeline) { + pc.bind( + 0, + &TraceData { + parameters: Parameters { + cam_position, + depth: FAR_Z, + cam_orientation: view.pose.orientation, + fov: [ + view.fov.angle_left, + view.fov.angle_right, + view.fov.angle_down, + view.fov.angle_up, + ], + torus_radius: TORUS_RADIUS, + rotation_angle, + pad: [0.0; 2], + }, + acc_struct: self.tlas, + output: self.target_view, + }, + ); + pc.dispatch(groups); + } } - } - context.sync_buffer(self.params_buf); - for eye in 0..view_count { - let eye_view = frame.xr_texture_view(eye); let mut pass = self.command_encoder.render( - "xr-eye", + "draw", gpu::RenderTargetSet { colors: &[gpu::RenderTarget { - view: eye_view, + view: frame.xr_texture_view(eye), init_op: gpu::InitOp::Clear(gpu::TextureColor::OpaqueBlack), finish_op: gpu::FinishOp::Store, }], - depth_stencil: Some(gpu::RenderTarget { - view: self.depth_views[eye as usize], - init_op: gpu::InitOp::Clear(gpu::TextureColor::White), - finish_op: gpu::FinishOp::Discard, - }), - }, - ); - let mut rc = pass.with(&mut self.pipeline); - let params_offset = (eye as usize * std::mem::size_of::()) as u64; - rc.bind( - 0, - &Params { - globals: self.params_buf.at(params_offset), + depth_stencil: None, }, ); - rc.draw(0, 12, 0, 1); + if let mut pc = pass.with(&self.draw_pipeline) { + pc.bind( + 0, + &DrawData { + input: self.target_view, + }, + ); + pc.draw(0, 3, 0, 1); + } } self.command_encoder.present(frame); - let _sync_point = context.submit(&mut self.command_encoder); + let sync_point = context.submit(&mut self.command_encoder); + + if let Some(sp) = self.prev_sync_point.take() { + context.wait_for(&sp, !0); + } + self.prev_sync_point = Some(sync_point); self.rendered_frames += 1; if self.xr_debug && (self.rendered_frames <= 5 || self.rendered_frames % 120 == 0) { info!("XR frame submitted: {}", self.rendered_frames); @@ -234,17 +346,17 @@ impl Example { } fn destroy(mut self, context: &gpu::Context) { - for view in &mut self.depth_views { - if *view != gpu::TextureView::default() { - context.destroy_texture_view(*view); - *view = gpu::TextureView::default(); - } + if let Some(sp) = self.prev_sync_point.take() { + context.wait_for(&sp, !0); } - context.destroy_texture(self.depth_texture); - context.destroy_buffer(self.params_buf); + context.destroy_texture_view(self.target_view); + context.destroy_texture(self.target); + context.destroy_acceleration_structure(self.blas); + context.destroy_acceleration_structure(self.tlas); context.destroy_xr_surface(&mut self.xr_surface); context.destroy_command_encoder(&mut self.command_encoder); - context.destroy_render_pipeline(&mut self.pipeline); + context.destroy_compute_pipeline(&mut self.rt_pipeline); + context.destroy_render_pipeline(&mut self.draw_pipeline); } } @@ -327,6 +439,7 @@ pub fn main() { .system(xr::FormFactor::HEAD_MOUNTED_DISPLAY) .unwrap(); mark!("XR mark: OpenXR system acquired"); + let context = unsafe { gpu::Context::init(gpu::ContextDesc { xr: Some(gpu::XrDesc { @@ -373,23 +486,12 @@ pub fn main() { } match e.state() { xr::SessionState::READY => { - if xr_debug { - info!("XR state READY -> begin session"); - } - mark!("XR mark: calling session.begin"); context.xr_session().unwrap().begin(VIEW_TYPE).unwrap(); - mark!("XR mark: session.begin returned"); - if matches!(state, AppState::Idle) { - mark!("XR mark: calling Example::new"); state = AppState::Running(Example::new(&context, xr_debug)); - mark!("XR mark: Example created"); } } xr::SessionState::STOPPING => { - if xr_debug { - info!("XR state STOPPING -> end session"); - } context.xr_session().unwrap().end().unwrap(); if let AppState::Running(example) = std::mem::replace(&mut state, AppState::Idle) @@ -398,20 +500,12 @@ pub fn main() { } } xr::SessionState::EXITING | xr::SessionState::LOSS_PENDING => { - if xr_debug { - info!("XR state {:?} -> exiting main loop", e.state()); - } break 'main_loop; } _ => {} } } - InstanceLossPending(_) => { - if xr_debug { - info!("XR instance loss pending -> exiting main loop"); - } - break 'main_loop; - } + InstanceLossPending(_) => break 'main_loop, _ => {} } } diff --git a/examples-android/xr.wgsl b/examples-android/xr.wgsl index f3e84333..01b280d1 100644 --- a/examples-android/xr.wgsl +++ b/examples-android/xr.wgsl @@ -1,52 +1,116 @@ -struct Params { - mvp: mat4x4, - tint: vec4, -}; -var globals: Params; +const MAX_BOUNCES: i32 = 3; -struct VsOut { - @builtin(position) position: vec4, - @location(0) color: vec4, +struct Parameters { + cam_position: vec3, + depth: f32, + cam_orientation: vec4, + fov: vec4, // left, right, down, up (radians) + torus_radius: f32, + rotation_angle: f32, }; -@vertex -fn vs_main(@builtin(vertex_index) vertex_index: u32) -> VsOut { - var positions = array, 12>( - vec3(0.0, 0.6, 0.0), - vec3(0.5, -0.3, 0.4), - vec3(-0.5, -0.3, 0.4), - vec3(0.0, 0.6, 0.0), - vec3(-0.5, -0.3, 0.4), - vec3(0.0, -0.3, -0.5), - vec3(0.0, 0.6, 0.0), - vec3(0.0, -0.3, -0.5), - vec3(0.5, -0.3, 0.4), - vec3(-0.5, -0.3, 0.4), - vec3(0.5, -0.3, 0.4), - vec3(0.0, -0.3, -0.5), +var parameters: Parameters; +var acc_struct: acceleration_structure; +var output: texture_storage_2d; + +fn qrot(q: vec4, v: vec3) -> vec3 { + return v + 2.0 * cross(q.xyz, cross(q.xyz, v) + q.w * v); +} + +fn qmake(axis: vec3, angle: f32) -> vec4 { + return vec4(axis * sin(angle), cos(angle)); +} + +fn get_miss_color(dir: vec3) -> vec4 { + var colors = array, 4>( + vec4(1.0), + vec4(0.6, 0.9, 0.3, 1.0), + vec4(0.3, 0.6, 0.9, 1.0), + vec4(0.0) ); - var colors = array, 12>( - vec4(1.0, 0.2, 0.2, 1.0), - vec4(0.2, 1.0, 0.2, 1.0), - vec4(0.2, 0.2, 1.0, 1.0), - vec4(1.0, 0.2, 0.2, 1.0), - vec4(0.2, 0.2, 1.0, 1.0), - vec4(1.0, 1.0, 0.2, 1.0), - vec4(1.0, 0.2, 0.2, 1.0), - vec4(1.0, 1.0, 0.2, 1.0), - vec4(0.2, 1.0, 0.2, 1.0), - vec4(0.2, 0.2, 1.0, 1.0), - vec4(0.2, 1.0, 0.2, 1.0), - vec4(1.0, 1.0, 0.2, 1.0), + var thresholds = array(-1.0, -0.3, 0.4, 1.0); + var i = 0; + loop { + if (dir.y < thresholds[i]) { + let t = (dir.y - thresholds[i - 1]) / (thresholds[i] - thresholds[i - 1]); + return mix(colors[i - 1], colors[i], t); + } + i += 1; + if (i >= 4) { + break; + } + } + return colors[3]; +} + +fn get_torus_normal(world_point: vec3, intersection: RayIntersection) -> vec3 { + // Match the original desktop ray-query example. + let local_point = intersection.world_to_object * vec4(world_point, 1.0); + let point_on_guiding_line = normalize(local_point.xy) * parameters.torus_radius; + let world_point_on_guiding_line = + intersection.object_to_world * vec4(point_on_guiding_line, 0.0, 1.0); + return normalize(world_point - world_point_on_guiding_line.xyz); +} + +@compute +@workgroup_size(8, 8) +fn main(@builtin(global_invocation_id) global_id: vec3) { + let target_size = textureDimensions(output); + if (any(global_id.xy >= target_size)) { + return; + } + + let uv = (vec2(global_id.xy) + vec2(0.5)) / vec2(target_size); + let tan_left = tan(parameters.fov.x); + let tan_right = tan(parameters.fov.y); + let tan_down = tan(parameters.fov.z); + let tan_up = tan(parameters.fov.w); + // uv.y=0 is top scanline, so map top->up and bottom->down. + let local_dir = vec3( + mix(tan_left, tan_right, uv.x), + mix(tan_up, tan_down, uv.y), + -1.0, ); + let world_dir = normalize(qrot(parameters.cam_orientation, local_dir)); + let rotator = qmake(vec3(0.0, 1.0, 0.0), parameters.rotation_angle); + var num_bounces = 0; + var rq: ray_query; + var ray_pos = qrot(rotator, parameters.cam_position); + var ray_dir = qrot(rotator, world_dir); + loop { + rayQueryInitialize( + &rq, + acc_struct, + RayDesc(RAY_FLAG_NONE, 0xFFu, 0.1, parameters.depth, ray_pos, ray_dir), + ); + rayQueryProceed(&rq); + let intersection = rayQueryGetCommittedIntersection(&rq); + if (intersection.kind == RAY_QUERY_INTERSECTION_NONE) { + break; + } + + ray_pos += ray_dir * intersection.t; + let normal = get_torus_normal(ray_pos, intersection); + ray_dir -= 2.0 * dot(ray_dir, normal) * normal; + + num_bounces += 1; + if (num_bounces > MAX_BOUNCES) { + break; + } + } - var out: VsOut; - out.position = globals.mvp * vec4(positions[vertex_index], 1.0); - out.color = colors[vertex_index]; - return out; + let color = get_miss_color(ray_dir); + textureStore(output, global_id.xy, color); } +@vertex +fn draw_vs(@builtin(vertex_index) vi: u32) -> @builtin(position) vec4 { + return vec4(f32(vi & 1u) * 4.0 - 1.0, f32(vi & 2u) * 2.0 - 1.0, 0.0, 1.0); +} + +var input: texture_2d; + @fragment -fn fs_main(in: VsOut) -> @location(0) vec4 { - return in.color * globals.tint; +fn draw_fs(@builtin(position) frag_coord: vec4) -> @location(0) vec4 { + return textureLoad(input, vec2(frag_coord.xy), 0); }