diff --git a/Cargo.lock b/Cargo.lock index 8e0f264d..34382207 100644 --- a/Cargo.lock +++ b/Cargo.lock @@ -838,6 +838,14 @@ name = "fuchsia-zircon-sys" version = "0.3.3" source = "registry+https://github.com/rust-lang/crates.io-index" +[[package]] +name = "fxhash" +version = "0.2.1" +source = "registry+https://github.com/rust-lang/crates.io-index" +dependencies = [ + "byteorder 1.3.4 (registry+https://github.com/rust-lang/crates.io-index)", +] + [[package]] name = "gcc" version = "0.3.55" @@ -1725,6 +1733,7 @@ dependencies = [ "bitflags 1.2.1 (registry+https://github.com/rust-lang/crates.io-index)", "byteorder 1.3.4 (registry+https://github.com/rust-lang/crates.io-index)", "crossbeam-channel 0.4.2 (registry+https://github.com/rust-lang/crates.io-index)", + "fxhash 0.2.1 (registry+https://github.com/rust-lang/crates.io-index)", "half 1.5.0 (registry+https://github.com/rust-lang/crates.io-index)", "hashbrown 0.7.1 (registry+https://github.com/rust-lang/crates.io-index)", "instant 0.1.2 (registry+https://github.com/rust-lang/crates.io-index)", @@ -2839,6 +2848,7 @@ source = "registry+https://github.com/rust-lang/crates.io-index" "checksum fs_extra 1.1.0 (registry+https://github.com/rust-lang/crates.io-index)" = "5f2a4a2034423744d2cc7ca2068453168dcdb82c438419e639a26bd87839c674" "checksum fuchsia-zircon 0.3.3 (registry+https://github.com/rust-lang/crates.io-index)" = "2e9763c69ebaae630ba35f74888db465e49e259ba1bc0eda7d06f4a067615d82" "checksum fuchsia-zircon-sys 0.3.3 (registry+https://github.com/rust-lang/crates.io-index)" = "3dcaa9ae7725d12cdb85b3ad99a434db70b468c09ded17e012d86b5c1010f7a7" +"checksum fxhash 0.2.1 (registry+https://github.com/rust-lang/crates.io-index)" = "c31b6d751ae2c7f11320402d34e41349dd1016f8d5d45e48c4312bc8625af50c" "checksum gcc 0.3.55 (registry+https://github.com/rust-lang/crates.io-index)" = "8f5f3913fa0bfe7ee1fd8248b6b9f42a5af4b9d65ec2dd2c3c26132b950ecfc2" "checksum getrandom 0.1.14 (registry+https://github.com/rust-lang/crates.io-index)" = "7abc8dd8451921606d809ba32e95b6111925cd2906060d2dcc29c070220503eb" "checksum gif 0.10.3 (registry+https://github.com/rust-lang/crates.io-index)" = "471d90201b3b223f3451cd4ad53e34295f16a1df17b1edf3736d47761c3981af" diff --git a/renderer/Cargo.toml b/renderer/Cargo.toml index da5e46b1..0bac653e 100644 --- a/renderer/Cargo.toml +++ b/renderer/Cargo.toml @@ -12,6 +12,7 @@ homepage = "https://github.com/servo/pathfinder" bitflags = "1.0" byteorder = "1.2" crossbeam-channel = "0.4" +fxhash = "0.2" half = "1.5" hashbrown = "0.7" rayon = "1.0" diff --git a/renderer/src/builder.rs b/renderer/src/builder.rs index 3724cbb6..cfebc416 100644 --- a/renderer/src/builder.rs +++ b/renderer/src/builder.rs @@ -12,8 +12,8 @@ use crate::concurrent::executor::Executor; use crate::gpu::renderer::{BlendModeExt, MASK_TILES_ACROSS, MASK_TILES_DOWN}; -use crate::gpu_data::{AlphaTileId, Fill, FillBatchEntry, RenderCommand, Tile, TileBatch}; -use crate::gpu_data::{TileBatchTexture, TileObjectPrimitive}; +use crate::gpu_data::{AlphaTileId, Clip, ClipBatch, ClipBatchKey, ClipBatchKind, Fill, FillBatchEntry, RenderCommand}; +use crate::gpu_data::{Tile, TileBatch, TileBatchTexture, TileObjectPrimitive}; use crate::options::{PreparedBuildOptions, PreparedRenderTransform, RenderCommandListener}; use crate::paint::{PaintInfo, PaintMetadata}; use crate::scene::{DisplayItem, Scene}; @@ -31,14 +31,17 @@ use pathfinder_geometry::util; use pathfinder_geometry::vector::{Vector2F, Vector2I, vec2f, vec2i}; use pathfinder_gpu::TextureSamplingFlags; use pathfinder_simd::default::{F32x4, I32x4}; -use std::sync::atomic::{AtomicUsize, Ordering}; +use std::sync::atomic::AtomicUsize; use instant::Instant; use std::u32; +pub(crate) const ALPHA_TILE_LEVEL_COUNT: usize = 2; +pub(crate) const ALPHA_TILES_PER_LEVEL: usize = 1 << (32 - ALPHA_TILE_LEVEL_COUNT + 1); + pub(crate) struct SceneBuilder<'a, 'b> { scene: &'a mut Scene, built_options: &'b PreparedBuildOptions, - next_alpha_tile_index: AtomicUsize, + next_alpha_tile_indices: [AtomicUsize; ALPHA_TILE_LEVEL_COUNT], pub(crate) listener: Box, } @@ -57,7 +60,6 @@ struct BuiltDrawPath { color_texture: Option, sampling_flags_1: TextureSamplingFlags, mask_0_fill_rule: FillRule, - mask_1_fill_rule: Option, } #[derive(Debug)] @@ -65,7 +67,7 @@ pub(crate) struct BuiltPath { pub solid_tiles: SolidTiles, pub empty_tiles: Vec, pub single_mask_tiles: Vec, - pub dual_mask_tiles: Vec, + pub clip_tiles: Vec, pub tiles: DenseTileMap, pub fill_rule: FillRule, } @@ -76,6 +78,12 @@ pub struct BuiltTile { pub tile: Tile, } +#[derive(Clone, Copy, Debug)] +pub struct BuiltClip { + pub clip: Clip, + pub key: ClipBatchKey, +} + #[derive(Clone, Debug)] pub(crate) enum SolidTiles { Occluders(Vec), @@ -96,7 +104,7 @@ impl<'a, 'b> SceneBuilder<'a, 'b> { SceneBuilder { scene, built_options, - next_alpha_tile_index: AtomicUsize::new(0), + next_alpha_tile_indices: [AtomicUsize::new(0), AtomicUsize::new(0)], listener, } } @@ -216,7 +224,6 @@ impl<'a, 'b> SceneBuilder<'a, 'b> { color_texture: paint_metadata.tile_batch_texture(), sampling_flags_1: TextureSamplingFlags::empty(), mask_0_fill_rule: path_object.fill_rule(), - mask_1_fill_rule: built_clip_path.map(|_| FillRule::Winding), } } @@ -226,6 +233,29 @@ impl<'a, 'b> SceneBuilder<'a, 'b> { } } + fn build_clips(&self, built_draw_paths: &[BuiltDrawPath]) { + let mut built_clip_tiles = vec![]; + for built_draw_path in built_draw_paths { + for built_clip_tile in &built_draw_path.path.clip_tiles { + built_clip_tiles.push(*built_clip_tile); + } + } + + built_clip_tiles.sort_by_key(|built_clip_tile| built_clip_tile.key); + + let mut batches: Vec = vec![]; + for built_clip_tile in built_clip_tiles { + if batches.is_empty() || batches.last_mut().unwrap().key != built_clip_tile.key { + batches.push(ClipBatch { key: built_clip_tile.key, clips: vec![] }); + } + batches.last_mut().unwrap().clips.push(built_clip_tile.clip); + } + + if !batches.is_empty() { + self.listener.send(RenderCommand::ClipTiles(batches)); + } + } + fn cull_tiles(&self, paint_metadata: &[PaintMetadata], built_draw_paths: Vec) -> CulledTiles { let mut culled_tiles = CulledTiles { display_list: vec![] }; @@ -280,7 +310,6 @@ impl<'a, 'b> SceneBuilder<'a, 'b> { None, built_draw_path.blend_mode, built_draw_path.filter, - None, None); self.add_alpha_tiles(&mut culled_tiles, @@ -290,20 +319,7 @@ impl<'a, 'b> SceneBuilder<'a, 'b> { color_texture, built_draw_path.blend_mode, built_draw_path.filter, - Some(built_draw_path.mask_0_fill_rule), - None); - - if let Some(mask_1_fill_rule) = built_draw_path.mask_1_fill_rule { - self.add_alpha_tiles(&mut culled_tiles, - layer_z_buffer, - &built_draw_path.path.dual_mask_tiles, - current_depth, - color_texture, - built_draw_path.blend_mode, - built_draw_path.filter, - Some(built_draw_path.mask_0_fill_rule), - Some(mask_1_fill_rule)); - } + Some(built_draw_path.mask_0_fill_rule)); match built_draw_path.path.solid_tiles { SolidTiles::Regular(ref tiles) => { @@ -314,7 +330,6 @@ impl<'a, 'b> SceneBuilder<'a, 'b> { color_texture, built_draw_path.blend_mode, built_draw_path.filter, - None, None); } SolidTiles::Occluders(_) => {} @@ -379,8 +394,7 @@ impl<'a, 'b> SceneBuilder<'a, 'b> { color_texture: Option, blend_mode: BlendMode, filter: Filter, - mask_0_fill_rule: Option, - mask_1_fill_rule: Option) { + mask_0_fill_rule: Option) { let mut batch_indices: Vec = vec![]; for built_alpha_tile in built_alpha_tiles { // Early cull if possible. @@ -406,13 +420,11 @@ impl<'a, 'b> SceneBuilder<'a, 'b> { blend_mode: batch_blend_mode, filter: batch_filter, mask_0_fill_rule: batch_mask_0_fill_rule, - mask_1_fill_rule: batch_mask_1_fill_rule, tile_page: batch_tile_page })) if *batch_color_texture == color_texture && batch_blend_mode == blend_mode && batch_filter == filter && batch_mask_0_fill_rule == mask_0_fill_rule && - batch_mask_1_fill_rule == mask_1_fill_rule && !batch_blend_mode.needs_readable_framebuffer() && batch_tile_page == built_alpha_tile.page => { dest_batch_index = Some(BatchIndex { @@ -438,7 +450,6 @@ impl<'a, 'b> SceneBuilder<'a, 'b> { blend_mode, filter, mask_0_fill_rule, - mask_1_fill_rule, tile_page: built_alpha_tile.page, })); } @@ -504,6 +515,7 @@ impl<'a, 'b> SceneBuilder<'a, 'b> { paint_metadata: &[PaintMetadata], built_draw_paths: Vec) { self.listener.send(RenderCommand::FlushFills); + self.build_clips(&built_draw_paths); let culled_tiles = self.cull_tiles(paint_metadata, built_draw_paths); self.pack_tiles(culled_tiles); } @@ -567,7 +579,7 @@ impl BuiltPath { BuiltPath { empty_tiles: vec![], single_mask_tiles: vec![], - dual_mask_tiles: vec![], + clip_tiles: vec![], solid_tiles: if occludes { SolidTiles::Occluders(vec![]) } else { @@ -688,9 +700,7 @@ impl ObjectBuilder { return alpha_tile_id; } - let alpha_tile_index = scene_builder.next_alpha_tile_index.fetch_add(1, Ordering::Relaxed); - debug_assert!(alpha_tile_index < u32::MAX as usize); - let alpha_tile_id = AlphaTileId(alpha_tile_index as u32); + let alpha_tile_id = AlphaTileId::new(&scene_builder.next_alpha_tile_indices, 0); self.built_path.tiles.data[local_tile_index].alpha_tile_id = alpha_tile_id; alpha_tile_id } @@ -799,31 +809,57 @@ impl ObjectBuilder { impl<'a> PackedTile<'a> { pub(crate) fn add_to(&self, tiles: &mut Vec, - draw_tiling_path_info: &DrawTilingPathInfo) { - let fill_tile_index = self.draw_tile.alpha_tile_id.tile() as u16; - let fill_tile_backdrop = self.draw_tile.backdrop as i8; - let (clip_tile_index, clip_tile_backdrop) = match self.clip_tile { - None => (0, 0), - Some(clip_tile) => { - // FIXME(pcwalton): This may not always be the case! - debug_assert!(!self.draw_tile.alpha_tile_id.is_valid() || - !clip_tile.alpha_tile_id.is_valid() || - self.draw_tile.alpha_tile_id.page() == - clip_tile.alpha_tile_id.page()); + clips: &mut Vec, + draw_tiling_path_info: &DrawTilingPathInfo, + scene_builder: &SceneBuilder) { + let draw_tile_page = self.draw_tile.alpha_tile_id.page() as u16; + let draw_tile_index = self.draw_tile.alpha_tile_id.tile() as u16; + let draw_tile_backdrop = self.draw_tile.backdrop as i8; - (clip_tile.alpha_tile_id.tile() as u16, clip_tile.backdrop as i8) + match self.clip_tile { + None => { + tiles.push(BuiltTile { + page: draw_tile_page, + tile: Tile::new_alpha(self.tile_coords, + draw_tile_index, + draw_tile_backdrop, + draw_tiling_path_info), + }); } - }; + Some(clip_tile) => { + let clip_tile_page = clip_tile.alpha_tile_id.page() as u16; + let clip_tile_index = clip_tile.alpha_tile_id.tile() as u16; + let clip_tile_backdrop = clip_tile.backdrop; - tiles.push(BuiltTile { - page: self.draw_tile.alpha_tile_id.page(), - tile: Tile::new_alpha(self.tile_coords, - fill_tile_index, - fill_tile_backdrop, - clip_tile_index, - clip_tile_backdrop, - draw_tiling_path_info), - }); + let dest_tile_id = AlphaTileId::new(&scene_builder.next_alpha_tile_indices, 1); + let dest_tile_page = dest_tile_id.page() as u16; + let dest_tile_index = dest_tile_id.tile() as u16; + + clips.push(BuiltClip { + clip: Clip::new(dest_tile_index, draw_tile_index, draw_tile_backdrop), + key: ClipBatchKey { + src_page: draw_tile_page, + dest_page: dest_tile_page, + kind: ClipBatchKind::Draw, + }, + }); + clips.push(BuiltClip { + clip: Clip::new(dest_tile_index, clip_tile_index, clip_tile_backdrop), + key: ClipBatchKey { + src_page: clip_tile_page, + dest_page: dest_tile_page, + kind: ClipBatchKind::Clip, + }, + }); + tiles.push(BuiltTile { + page: dest_tile_page, + tile: Tile::new_alpha(self.tile_coords, + dest_tile_index, + 0, + draw_tiling_path_info), + }); + } + } } } @@ -832,21 +868,17 @@ impl Tile { fn new_alpha(tile_origin: Vector2I, draw_tile_index: u16, draw_tile_backdrop: i8, - clip_tile_index: u16, - clip_tile_backdrop: i8, draw_tiling_path_info: &DrawTilingPathInfo) -> Tile { let mask_0_uv = calculate_mask_uv(draw_tile_index); - let mask_1_uv = calculate_mask_uv(clip_tile_index); Tile { tile_x: tile_origin.x() as i16, tile_y: tile_origin.y() as i16, mask_0_u: mask_0_uv.x() as u8, mask_0_v: mask_0_uv.y() as u8, - mask_1_u: mask_1_uv.x() as u8, - mask_1_v: mask_1_uv.y() as u8, mask_0_backdrop: draw_tile_backdrop, - mask_1_backdrop: clip_tile_backdrop, + flags: 0, + pad: 0, color: draw_tiling_path_info.paint_id.0, } } @@ -857,6 +889,23 @@ impl Tile { } } +impl Clip { + #[inline] + fn new(dest_tile_index: u16, src_tile_index: u16, src_backdrop: i8) -> Clip { + let dest_uv = calculate_mask_uv(dest_tile_index); + let src_uv = calculate_mask_uv(src_tile_index); + Clip { + dest_u: dest_uv.x() as u8, + dest_v: dest_uv.y() as u8, + src_u: src_uv.x() as u8, + src_v: src_uv.y() as u8, + backdrop: src_backdrop, + pad_0: 0, + pad_1: 0, + } + } +} + fn calculate_mask_uv(tile_index: u16) -> Vector2I { debug_assert_eq!(MASK_TILES_ACROSS, MASK_TILES_DOWN); let mask_u = tile_index as i32 % MASK_TILES_ACROSS as i32; diff --git a/renderer/src/gpu/renderer.rs b/renderer/src/gpu/renderer.rs index 29ebd585..4d9ae121 100644 --- a/renderer/src/gpu/renderer.rs +++ b/renderer/src/gpu/renderer.rs @@ -10,16 +10,17 @@ use crate::gpu::debug::DebugUIPresenter; use crate::gpu::options::{DestFramebuffer, RendererOptions}; -use crate::gpu::shaders::{BlitProgram, BlitVertexArray, CopyTileProgram, CopyTileVertexArray}; -use crate::gpu::shaders::{FillProgram, FillVertexArray, MAX_FILLS_PER_BATCH, ReprojectionProgram}; -use crate::gpu::shaders::{ReprojectionVertexArray, StencilProgram, StencilVertexArray}; -use crate::gpu::shaders::{TileProgram, TileVertexArray}; -use crate::gpu_data::{Fill, FillBatchEntry, RenderCommand, TextureLocation}; -use crate::gpu_data::{TextureMetadataEntry, TexturePageDescriptor, TexturePageId}; +use crate::gpu::shaders::{BlitProgram, BlitVertexArray, ClipTileProgram, ClipTileVertexArray}; +use crate::gpu::shaders::{CopyTileProgram, CopyTileVertexArray, FillProgram, FillVertexArray}; +use crate::gpu::shaders::{MAX_FILLS_PER_BATCH, ReprojectionProgram, ReprojectionVertexArray}; +use crate::gpu::shaders::{StencilProgram, StencilVertexArray, TileProgram, TileVertexArray}; +use crate::gpu_data::{ClipBatch, ClipBatchKey, ClipBatchKind, Fill, FillBatchEntry, RenderCommand}; +use crate::gpu_data::{TextureLocation, TextureMetadataEntry, TexturePageDescriptor, TexturePageId}; use crate::gpu_data::{Tile, TileBatchTexture}; use crate::options::BoundingQuad; use crate::paint::PaintCompositeOp; use crate::tiles::{TILE_HEIGHT, TILE_WIDTH}; +use fxhash::FxHashMap; use half::f16; use pathfinder_color::{self as color, ColorF, ColorU}; use pathfinder_content::effects::{BlendMode, BlurDirection, DefringingKernel}; @@ -31,7 +32,7 @@ use pathfinder_geometry::rect::RectI; use pathfinder_geometry::transform3d::Transform4F; use pathfinder_geometry::util; use pathfinder_geometry::vector::{Vector2F, Vector2I, Vector4F, vec2f, vec2i}; -use pathfinder_gpu::{BlendFactor, BlendState, BufferData, BufferTarget, BufferUploadMode}; +use pathfinder_gpu::{BlendFactor, BlendOp, BlendState, BufferData, BufferTarget, BufferUploadMode}; use pathfinder_gpu::{ClearOps, DepthFunc, DepthState, Device, Primitive, RenderOptions}; use pathfinder_gpu::{RenderState, RenderTarget, StencilFunc, StencilState, TextureDataRef}; use pathfinder_gpu::{TextureFormat, UniformData}; @@ -92,7 +93,6 @@ const COMBINER_CTRL_COMPOSITE_COLOR: i32 = 0xe; const COMBINER_CTRL_COMPOSITE_LUMINOSITY: i32 = 0xf; const COMBINER_CTRL_MASK_0_SHIFT: i32 = 0; -const COMBINER_CTRL_MASK_1_SHIFT: i32 = 2; const COMBINER_CTRL_COLOR_FILTER_SHIFT: i32 = 4; const COMBINER_CTRL_COLOR_COMBINE_SHIFT: i32 = 6; const COMBINER_CTRL_COMPOSITE_SHIFT: i32 = 8; @@ -111,16 +111,18 @@ where fill_program: FillProgram, tile_program: TileProgram, tile_copy_program: CopyTileProgram, + tile_clip_program: ClipTileProgram, blit_vertex_array: BlitVertexArray, tile_vertex_array: TileVertexArray, tile_copy_vertex_array: CopyTileVertexArray, + tile_clip_vertex_array: ClipTileVertexArray, tile_vertex_buffer: D::Buffer, quad_vertex_positions_buffer: D::Buffer, quad_vertex_indices_buffer: D::Buffer, quads_vertex_indices_buffer: D::Buffer, quads_vertex_indices_length: usize, fill_vertex_array: FillVertexArray, - alpha_tile_pages: Vec>, + alpha_tile_pages: FxHashMap>, dest_blend_framebuffer: D::Framebuffer, intermediate_dest_framebuffer: D::Framebuffer, texture_pages: Vec>>, @@ -167,6 +169,7 @@ where let fill_program = FillProgram::new(&device, resources); let tile_program = TileProgram::new(&device, resources); let tile_copy_program = CopyTileProgram::new(&device, resources); + let tile_clip_program = ClipTileProgram::new(&device, resources); let stencil_program = StencilProgram::new(&device, resources); let reprojection_program = ReprojectionProgram::new(&device, resources); @@ -219,6 +222,12 @@ where &tile_vertex_buffer, &quads_vertex_indices_buffer, ); + let tile_clip_vertex_array = ClipTileVertexArray::new( + &device, + &tile_clip_program, + &quad_vertex_positions_buffer, + &quad_vertex_indices_buffer, + ); let stencil_vertex_array = StencilVertexArray::new(&device, &stencil_program); let reprojection_vertex_array = ReprojectionVertexArray::new( &device, @@ -249,16 +258,18 @@ where fill_program, tile_program, tile_copy_program, + tile_clip_program, blit_vertex_array, tile_vertex_array, tile_copy_vertex_array, + tile_clip_vertex_array, tile_vertex_buffer, quad_vertex_positions_buffer, quad_vertex_indices_buffer, quads_vertex_indices_buffer, quads_vertex_indices_length: 0, fill_vertex_array, - alpha_tile_pages: vec![], + alpha_tile_pages: FxHashMap::default(), dest_blend_framebuffer, intermediate_dest_framebuffer, texture_pages: vec![], @@ -291,7 +302,7 @@ where pub fn begin_scene(&mut self) { self.framebuffer_flags = FramebufferFlags::empty(); - for alpha_tile_page in &mut self.alpha_tile_pages { + for alpha_tile_page in self.alpha_tile_pages.values_mut() { alpha_tile_page.must_preserve_framebuffer = false; } @@ -318,10 +329,14 @@ where } RenderCommand::AddFills(ref fills) => self.add_fills(fills), RenderCommand::FlushFills => { - for page_index in 0..(self.alpha_tile_pages.len() as u16) { + let page_indices: Vec<_> = self.alpha_tile_pages.keys().cloned().collect(); + for page_index in page_indices { self.draw_buffered_fills(page_index) } } + RenderCommand::ClipTiles(ref batches) => { + batches.iter().for_each(|batch| self.draw_clip_batch(batch)) + } RenderCommand::BeginTileDrawing => self.begin_tile_drawing(), RenderCommand::PushRenderTarget(render_target_id) => { self.push_render_target(render_target_id) @@ -335,7 +350,6 @@ where count as u32, batch.color_texture, batch.mask_0_fill_rule, - batch.mask_1_fill_rule, batch.blend_mode, batch.filter) } @@ -571,20 +585,26 @@ where for fill_batch_entry in fill_batch { let page = fill_batch_entry.page; - while self.alpha_tile_pages.len() <= page as usize { - self.alpha_tile_pages.push(AlphaTilePage::new(&mut self.device)); + if !self.alpha_tile_pages.contains_key(&page) { + self.alpha_tile_pages.insert(page, AlphaTilePage::new(&mut self.device)); } - self.alpha_tile_pages[page as usize].buffered_fills.push(fill_batch_entry.fill); - if self.alpha_tile_pages[page as usize].buffered_fills.len() == MAX_FILLS_PER_BATCH { + if self.alpha_tile_pages[&page].buffered_fills.len() == MAX_FILLS_PER_BATCH { self.draw_buffered_fills(page); } + self.alpha_tile_pages + .get_mut(&page) + .unwrap() + .buffered_fills + .push(fill_batch_entry.fill); } } fn draw_buffered_fills(&mut self, page: u16) { let mask_viewport = self.mask_viewport(); - let alpha_tile_page = &mut self.alpha_tile_pages[page as usize]; + let alpha_tile_page = self.alpha_tile_pages + .get_mut(&page) + .expect("Where's the alpha tile page?"); let buffered_fills = &mut alpha_tile_page.buffered_fills; if buffered_fills.is_empty() { return; @@ -635,6 +655,67 @@ where buffered_fills.clear(); } + fn draw_clip_batch(&mut self, batch: &ClipBatch) { + if batch.clips.is_empty() { + return; + } + + let ClipBatchKey { dest_page, src_page, kind } = batch.key; + + self.device.allocate_buffer(&self.tile_clip_vertex_array.vertex_buffer, + BufferData::Memory(&batch.clips), + BufferTarget::Vertex, + BufferUploadMode::Dynamic); + + if !self.alpha_tile_pages.contains_key(&dest_page) { + self.alpha_tile_pages.insert(dest_page, AlphaTilePage::new(&mut self.device)); + } + + let mut clear_color = None; + if !self.alpha_tile_pages[&dest_page].must_preserve_framebuffer { + clear_color = Some(ColorF::default()); + }; + + let blend = match kind { + ClipBatchKind::Draw => None, + ClipBatchKind::Clip => { + Some(BlendState { + src_rgb_factor: BlendFactor::One, + src_alpha_factor: BlendFactor::One, + dest_rgb_factor: BlendFactor::One, + dest_alpha_factor: BlendFactor::One, + op: BlendOp::Min, + }) + } + }; + + let mask_viewport = self.mask_viewport(); + + { + let dest_framebuffer = &self.alpha_tile_pages[&dest_page].framebuffer; + let src_framebuffer = &self.alpha_tile_pages[&src_page].framebuffer; + let src_texture = self.device.framebuffer_texture(&src_framebuffer); + + debug_assert!(batch.clips.len() <= u32::MAX as usize); + self.device.draw_elements_instanced(6, batch.clips.len() as u32, &RenderState { + target: &RenderTarget::Framebuffer(dest_framebuffer), + program: &self.tile_clip_program.program, + vertex_array: &self.tile_clip_vertex_array.vertex_array, + primitive: Primitive::Triangles, + textures: &[src_texture], + uniforms: &[(&self.tile_clip_program.src_uniform, UniformData::TextureUnit(0))], + viewport: mask_viewport, + options: RenderOptions { + blend, + clear_ops: ClearOps { color: clear_color, ..ClearOps::default() }, + ..RenderOptions::default() + }, + }); + } + + self.alpha_tile_pages.get_mut(&dest_page).unwrap().must_preserve_framebuffer = true; + } + fn tile_transform(&self) -> Transform4F { let draw_viewport = self.draw_viewport().size().to_f32(); let scale = Vector4F::new(2.0 / draw_viewport.x(), -2.0 / draw_viewport.y(), 1.0, 1.0); @@ -646,7 +727,6 @@ where tile_count: u32, color_texture_0: Option, mask_0_fill_rule: Option, - mask_1_fill_rule: Option, blend_mode: BlendMode, filter: Filter) { // TODO(pcwalton): Disable blend for solid tiles. @@ -660,14 +740,13 @@ where let draw_viewport = self.draw_viewport(); let mut ctrl = 0; - for &(fill_rule, shift) in &[ - (mask_0_fill_rule, COMBINER_CTRL_MASK_0_SHIFT), - (mask_1_fill_rule, COMBINER_CTRL_MASK_1_SHIFT), - ] { - match fill_rule { - None => {} - Some(FillRule::Winding) => ctrl |= COMBINER_CTRL_MASK_WINDING << shift, - Some(FillRule::EvenOdd) => ctrl |= COMBINER_CTRL_MASK_EVEN_ODD << shift, + match mask_0_fill_rule { + None => {} + Some(FillRule::Winding) => { + ctrl |= COMBINER_CTRL_MASK_WINDING << COMBINER_CTRL_MASK_0_SHIFT + } + Some(FillRule::EvenOdd) => { + ctrl |= COMBINER_CTRL_MASK_EVEN_ODD << COMBINER_CTRL_MASK_0_SHIFT } } @@ -695,13 +774,7 @@ where uniforms.push((&self.tile_program.mask_texture_0_uniform, UniformData::TextureUnit(textures.len() as u32))); textures.push(self.device.framebuffer_texture( - &self.alpha_tile_pages[tile_page as usize].framebuffer)); - } - if mask_1_fill_rule.is_some() { - uniforms.push((&self.tile_program.mask_texture_1_uniform, - UniformData::TextureUnit(textures.len() as u32))); - textures.push(self.device.framebuffer_texture( - &self.alpha_tile_pages[tile_page as usize].framebuffer)); + &self.alpha_tile_pages[&tile_page].framebuffer)); } // TODO(pcwalton): Refactor. diff --git a/renderer/src/gpu/shaders.rs b/renderer/src/gpu/shaders.rs index c3e9f6e9..39268183 100644 --- a/renderer/src/gpu/shaders.rs +++ b/renderer/src/gpu/shaders.rs @@ -16,6 +16,7 @@ use pathfinder_resources::ResourceLoader; // TODO(pcwalton): Replace with `mem::size_of` calls? const FILL_INSTANCE_SIZE: usize = 8; const TILE_INSTANCE_SIZE: usize = 12; +const CLIP_TILE_INSTANCE_SIZE: usize = 8; pub const MAX_FILLS_PER_BATCH: usize = 0x4000; @@ -165,8 +166,6 @@ impl TileVertexArray where D: Device { device.get_vertex_attr(&tile_program.program, "TileOrigin").unwrap(); let mask_0_tex_coord_attr = device.get_vertex_attr(&tile_program.program, "MaskTexCoord0").unwrap(); - let mask_1_tex_coord_attr = - device.get_vertex_attr(&tile_program.program, "MaskTexCoord1").unwrap(); let mask_backdrop_attr = device.get_vertex_attr(&tile_program.program, "MaskBackdrop").unwrap(); let color_attr = device.get_vertex_attr(&tile_program.program, "Color").unwrap(); @@ -200,21 +199,12 @@ impl TileVertexArray where D: Device { divisor: 1, buffer_index: 1, }); - device.configure_vertex_attr(&vertex_array, &mask_1_tex_coord_attr, &VertexAttrDescriptor { - size: 2, - class: VertexAttrClass::Int, - attr_type: VertexAttrType::U8, - stride: TILE_INSTANCE_SIZE, - offset: 6, - divisor: 1, - buffer_index: 1, - }); device.configure_vertex_attr(&vertex_array, &mask_backdrop_attr, &VertexAttrDescriptor { size: 2, class: VertexAttrClass::Int, attr_type: VertexAttrType::I8, stride: TILE_INSTANCE_SIZE, - offset: 8, + offset: 6, divisor: 1, buffer_index: 1, }); @@ -223,7 +213,7 @@ impl TileVertexArray where D: Device { class: VertexAttrClass::Int, attr_type: VertexAttrType::I16, stride: TILE_INSTANCE_SIZE, - offset: 10, + offset: 8, divisor: 1, buffer_index: 1, }); @@ -265,6 +255,74 @@ impl CopyTileVertexArray where D: Device { } } +pub struct ClipTileVertexArray where D: Device { + pub vertex_array: D::VertexArray, + pub vertex_buffer: D::Buffer, +} + +impl ClipTileVertexArray where D: Device { + pub fn new(device: &D, + clip_tile_program: &ClipTileProgram, + quad_vertex_positions_buffer: &D::Buffer, + quad_vertex_indices_buffer: &D::Buffer) + -> ClipTileVertexArray { + let vertex_array = device.create_vertex_array(); + let vertex_buffer = device.create_buffer(); + + let tile_offset_attr = + device.get_vertex_attr(&clip_tile_program.program, "TileOffset").unwrap(); + let dest_tile_origin_attr = + device.get_vertex_attr(&clip_tile_program.program, "DestTileOrigin").unwrap(); + let src_tile_origin_attr = + device.get_vertex_attr(&clip_tile_program.program, "SrcTileOrigin").unwrap(); + let src_backdrop_attr = + device.get_vertex_attr(&clip_tile_program.program, "SrcBackdrop").unwrap(); + + device.bind_buffer(&vertex_array, quad_vertex_positions_buffer, BufferTarget::Vertex); + device.configure_vertex_attr(&vertex_array, &tile_offset_attr, &VertexAttrDescriptor { + size: 2, + class: VertexAttrClass::Int, + attr_type: VertexAttrType::I16, + stride: 4, + offset: 0, + divisor: 0, + buffer_index: 0, + }); + device.bind_buffer(&vertex_array, &vertex_buffer, BufferTarget::Vertex); + device.configure_vertex_attr(&vertex_array, &dest_tile_origin_attr, &VertexAttrDescriptor { + size: 2, + class: VertexAttrClass::Int, + attr_type: VertexAttrType::U8, + stride: CLIP_TILE_INSTANCE_SIZE, + offset: 0, + divisor: 1, + buffer_index: 1, + }); + device.configure_vertex_attr(&vertex_array, &src_tile_origin_attr, &VertexAttrDescriptor { + size: 2, + class: VertexAttrClass::Int, + attr_type: VertexAttrType::U8, + stride: CLIP_TILE_INSTANCE_SIZE, + offset: 2, + divisor: 1, + buffer_index: 1, + }); + device.configure_vertex_attr(&vertex_array, &src_backdrop_attr, &VertexAttrDescriptor { + size: 1, + class: VertexAttrClass::Int, + attr_type: VertexAttrType::I8, + stride: CLIP_TILE_INSTANCE_SIZE, + offset: 4, + divisor: 1, + buffer_index: 1, + }); + device.bind_buffer(&vertex_array, quad_vertex_indices_buffer, BufferTarget::Index); + + ClipTileVertexArray { vertex_array, vertex_buffer } + } +} + + pub struct BlitProgram where D: Device { pub program: D::Program, pub src_uniform: D::Uniform, @@ -316,7 +374,6 @@ pub struct TileProgram where D: Device { pub color_texture_0_uniform: D::Uniform, pub color_texture_1_uniform: D::Uniform, pub mask_texture_0_uniform: D::Uniform, - pub mask_texture_1_uniform: D::Uniform, pub gamma_lut_uniform: D::Uniform, pub color_texture_0_size_uniform: D::Uniform, pub filter_params_0_uniform: D::Uniform, @@ -337,7 +394,6 @@ impl TileProgram where D: Device { let color_texture_0_uniform = device.get_uniform(&program, "ColorTexture0"); let color_texture_1_uniform = device.get_uniform(&program, "ColorTexture1"); let mask_texture_0_uniform = device.get_uniform(&program, "MaskTexture0"); - let mask_texture_1_uniform = device.get_uniform(&program, "MaskTexture1"); let gamma_lut_uniform = device.get_uniform(&program, "GammaLUT"); let color_texture_0_size_uniform = device.get_uniform(&program, "ColorTexture0Size"); let filter_params_0_uniform = device.get_uniform(&program, "FilterParams0"); @@ -355,7 +411,6 @@ impl TileProgram where D: Device { color_texture_0_uniform, color_texture_1_uniform, mask_texture_0_uniform, - mask_texture_1_uniform, gamma_lut_uniform, color_texture_0_size_uniform, filter_params_0_uniform, @@ -392,6 +447,19 @@ impl CopyTileProgram where D: Device { } } +pub struct ClipTileProgram where D: Device { + pub program: D::Program, + pub src_uniform: D::Uniform, +} + +impl ClipTileProgram where D: Device { + pub fn new(device: &D, resources: &dyn ResourceLoader) -> ClipTileProgram { + let program = device.create_program(resources, "tile_clip"); + let src_uniform = device.get_uniform(&program, "Src"); + ClipTileProgram { program, src_uniform } + } +} + pub struct StencilProgram where D: Device, diff --git a/renderer/src/gpu_data.rs b/renderer/src/gpu_data.rs index cfc43cf5..7f9b0eeb 100644 --- a/renderer/src/gpu_data.rs +++ b/renderer/src/gpu_data.rs @@ -10,6 +10,7 @@ //! Packed data ready to be sent to the GPU. +use crate::builder::{ALPHA_TILES_PER_LEVEL, ALPHA_TILE_LEVEL_COUNT}; use crate::options::BoundingQuad; use crate::paint::PaintCompositeOp; use pathfinder_color::ColorU; @@ -23,7 +24,9 @@ use pathfinder_geometry::vector::Vector2I; use pathfinder_gpu::TextureSamplingFlags; use std::fmt::{Debug, Formatter, Result as DebugResult}; use std::sync::Arc; +use std::sync::atomic::{AtomicUsize, Ordering}; use std::time::Duration; +use std::u32; pub enum RenderCommand { // Starts rendering a frame. @@ -61,6 +64,9 @@ pub enum RenderCommand { // Flushes the queue of fills. FlushFills, + // Renders clips to the mask tile. + ClipTiles(Vec), + // Pushes a render target onto the stack. Draw commands go to the render target on top of the // stack. PushRenderTarget(RenderTargetId), @@ -97,7 +103,6 @@ pub struct TileBatch { pub tiles: Vec, pub color_texture: Option, pub mask_0_fill_rule: Option, - pub mask_1_fill_rule: Option, pub filter: Filter, pub blend_mode: BlendMode, pub tile_page: u16, @@ -147,6 +152,38 @@ pub struct Fill { pub alpha_tile_index: u16, } +#[derive(Clone, Debug)] +pub struct ClipBatch { + pub clips: Vec, + pub key: ClipBatchKey, +} + +#[derive(Clone, Copy, Debug, PartialEq, Eq, PartialOrd, Ord)] +pub struct ClipBatchKey { + pub dest_page: u16, + pub src_page: u16, + pub kind: ClipBatchKind, +} + +// Order is significant here. +#[derive(Clone, Copy, Debug, PartialEq, Eq, PartialOrd, Ord)] +pub enum ClipBatchKind { + Draw, + Clip, +} + +#[derive(Clone, Copy, Debug, Default)] +#[repr(C)] +pub struct Clip { + pub dest_u: u8, + pub dest_v: u8, + pub src_u: u8, + pub src_v: u8, + pub backdrop: i8, + pub pad_0: u8, + pub pad_1: u16, +} + #[derive(Clone, Copy, Debug, Default)] #[repr(C)] pub struct Tile { @@ -154,17 +191,24 @@ pub struct Tile { pub tile_y: i16, pub mask_0_u: u8, pub mask_0_v: u8, - pub mask_1_u: u8, - pub mask_1_v: u8, pub mask_0_backdrop: i8, - pub mask_1_backdrop: i8, + pub pad: u8, pub color: u16, + pub flags: u16, } #[derive(Clone, Copy, PartialEq, Debug)] pub struct AlphaTileId(pub u32); impl AlphaTileId { + #[inline] + pub fn new(next_alpha_tile_index: &[AtomicUsize; ALPHA_TILE_LEVEL_COUNT], level: usize) + -> AlphaTileId { + let alpha_tile_index = next_alpha_tile_index[level].fetch_add(1, Ordering::Relaxed); + debug_assert!(alpha_tile_index < ALPHA_TILES_PER_LEVEL); + AlphaTileId((level * ALPHA_TILES_PER_LEVEL + alpha_tile_index) as u32) + } + #[inline] pub fn invalid() -> AlphaTileId { AlphaTileId(!0) @@ -206,6 +250,9 @@ impl Debug for RenderCommand { write!(formatter, "AddFills(x{})", fills.len()) } RenderCommand::FlushFills => write!(formatter, "FlushFills"), + RenderCommand::ClipTiles(ref batches) => { + write!(formatter, "ClipTiles(x{})", batches.len()) + } RenderCommand::PushRenderTarget(render_target_id) => { write!(formatter, "PushRenderTarget({:?})", render_target_id) } diff --git a/renderer/src/tiles.rs b/renderer/src/tiles.rs index 36b35f23..b473a88b 100644 --- a/renderer/src/tiles.rs +++ b/renderer/src/tiles.rs @@ -146,23 +146,25 @@ impl<'a, 'b> Tiler<'a, 'b> { occluders.push(Occluder::new(packed_tile.tile_coords)); } SolidTiles::Regular(ref mut solid_tiles) => { - packed_tile.add_to(solid_tiles, &draw_tiling_path_info); + packed_tile.add_to(solid_tiles, + &mut self.object_builder.built_path.clip_tiles, + &draw_tiling_path_info, + &self.scene_builder); } } } TileType::SingleMask => { debug_assert_ne!(packed_tile.draw_tile.alpha_tile_id.page(), !0); packed_tile.add_to(&mut self.object_builder.built_path.single_mask_tiles, - &draw_tiling_path_info); - } - TileType::DualMask => { - debug_assert_ne!(packed_tile.draw_tile.alpha_tile_id.page(), !0); - packed_tile.add_to(&mut self.object_builder.built_path.dual_mask_tiles, - &draw_tiling_path_info); + &mut self.object_builder.built_path.clip_tiles, + &draw_tiling_path_info, + &self.scene_builder); } TileType::Empty if blend_mode_is_destructive => { packed_tile.add_to(&mut self.object_builder.built_path.empty_tiles, - &draw_tiling_path_info); + &mut self.object_builder.built_path.clip_tiles, + &draw_tiling_path_info, + &self.scene_builder); } TileType::Empty => { // Just cull. @@ -404,7 +406,6 @@ pub(crate) enum TileType { Solid, Empty, SingleMask, - DualMask, } impl<'a> PackedTile<'a> { @@ -504,7 +505,7 @@ impl<'a> PackedTile<'a> { Some(clip_tile) => { // We have both a draw and clip mask. Composite them together. PackedTile { - tile_type: TileType::DualMask, + tile_type: TileType::SingleMask, tile_coords, draw_tile, clip_tile: Some(clip_tile), diff --git a/renderer/src/z_buffer.rs b/renderer/src/z_buffer.rs index dfdc3413..e45889d9 100644 --- a/renderer/src/z_buffer.rs +++ b/renderer/src/z_buffer.rs @@ -92,7 +92,6 @@ impl ZBuffer { filter: Filter::None, blend_mode: BlendMode::default(), mask_0_fill_rule: None, - mask_1_fill_rule: None, tile_page: !0, }); } @@ -112,11 +111,10 @@ impl Tile { tile_x: tile_origin.x() as i16, tile_y: tile_origin.y() as i16, mask_0_backdrop: 0, - mask_1_backdrop: 0, mask_0_u: 0, mask_0_v: 0, - mask_1_u: 0, - mask_1_v: 0, + flags: 0, + pad: 0, color: paint_id.0, } } diff --git a/resources/shaders/gl3/tile.fs.glsl b/resources/shaders/gl3/tile.fs.glsl index e5202a04..cc9f530b 100644 --- a/resources/shaders/gl3/tile.fs.glsl +++ b/resources/shaders/gl3/tile.fs.glsl @@ -77,12 +77,10 @@ precision highp sampler2D; - uniform sampler2D uColorTexture0; uniform sampler2D uMaskTexture0; -uniform sampler2D uMaskTexture1; uniform sampler2D uDestTexture; uniform sampler2D uGammaLUT; uniform vec4 uFilterParams0; @@ -93,7 +91,6 @@ uniform vec2 uColorTexture0Size; uniform int uCtrl; in vec3 vMaskTexCoord0; -in vec3 vMaskTexCoord1; in vec2 vColorTexCoord0; in vec4 vBaseColor; @@ -562,10 +559,8 @@ float sampleMask(float maskAlpha, void calculateColor(int ctrl){ int maskCtrl0 =(ctrl >> 0)& 0x3; - int maskCtrl1 =(ctrl >> 2)& 0x3; float maskAlpha = 1.0; maskAlpha = sampleMask(maskAlpha, uMaskTexture0, vMaskTexCoord0, maskCtrl0); - maskAlpha = sampleMask(maskAlpha, uMaskTexture1, vMaskTexCoord1, maskCtrl1); vec4 color = vBaseColor; diff --git a/resources/shaders/gl3/tile.vs.glsl b/resources/shaders/gl3/tile.vs.glsl index 320414e8..23c67fee 100644 --- a/resources/shaders/gl3/tile.vs.glsl +++ b/resources/shaders/gl3/tile.vs.glsl @@ -23,12 +23,10 @@ uniform ivec2 uTextureMetadataSize; in ivec2 aTileOffset; in ivec2 aTileOrigin; in uvec2 aMaskTexCoord0; -in uvec2 aMaskTexCoord1; in ivec2 aMaskBackdrop; in int aColor; out vec3 vMaskTexCoord0; -out vec3 vMaskTexCoord1; out vec2 vColorTexCoord0; out vec4 vBaseColor; @@ -37,7 +35,6 @@ void main(){ vec2 position =(tileOrigin + tileOffset)* uTileSize; vec2 maskTexCoord0 =(vec2(aMaskTexCoord0)+ tileOffset)/ 256.0; - vec2 maskTexCoord1 =(vec2(aMaskTexCoord1)+ tileOffset)/ 256.0; vec2 textureMetadataScale = vec2(1.0)/ vec2(uTextureMetadataSize); vec2 metadataEntryCoord = vec2(aColor % 128 * 4, aColor / 128); @@ -50,7 +47,6 @@ void main(){ vColorTexCoord0 = mat2(colorTexMatrix0)* position + colorTexOffsets . xy; vMaskTexCoord0 = vec3(maskTexCoord0, float(aMaskBackdrop . x)); - vMaskTexCoord1 = vec3(maskTexCoord1, float(aMaskBackdrop . y)); vBaseColor = baseColor; gl_Position = uTransform * vec4(position, 0.0, 1.0); } diff --git a/resources/shaders/metal/tile.fs.metal b/resources/shaders/metal/tile.fs.metal index b422004a..89516bed 100644 --- a/resources/shaders/metal/tile.fs.metal +++ b/resources/shaders/metal/tile.fs.metal @@ -10,20 +10,18 @@ struct spvDescriptorSetBuffer0 { texture2d uMaskTexture0 [[id(0)]]; sampler uMaskTexture0Smplr [[id(1)]]; - texture2d uMaskTexture1 [[id(2)]]; - sampler uMaskTexture1Smplr [[id(3)]]; - texture2d uColorTexture0 [[id(4)]]; - sampler uColorTexture0Smplr [[id(5)]]; - texture2d uGammaLUT [[id(6)]]; - sampler uGammaLUTSmplr [[id(7)]]; - constant float2* uColorTexture0Size [[id(8)]]; - constant float2* uFramebufferSize [[id(9)]]; - constant float4* uFilterParams0 [[id(10)]]; - constant float4* uFilterParams1 [[id(11)]]; - constant float4* uFilterParams2 [[id(12)]]; - texture2d uDestTexture [[id(13)]]; - sampler uDestTextureSmplr [[id(14)]]; - constant int* uCtrl [[id(15)]]; + texture2d uColorTexture0 [[id(2)]]; + sampler uColorTexture0Smplr [[id(3)]]; + texture2d uGammaLUT [[id(4)]]; + sampler uGammaLUTSmplr [[id(5)]]; + constant float2* uColorTexture0Size [[id(6)]]; + constant float2* uFramebufferSize [[id(7)]]; + constant float4* uFilterParams0 [[id(8)]]; + constant float4* uFilterParams1 [[id(9)]]; + constant float4* uFilterParams2 [[id(10)]]; + texture2d uDestTexture [[id(11)]]; + sampler uDestTextureSmplr [[id(12)]]; + constant int* uCtrl [[id(13)]]; }; constant float3 _1040 = {}; @@ -36,9 +34,8 @@ struct main0_out struct main0_in { float3 vMaskTexCoord0 [[user(locn0)]]; - float3 vMaskTexCoord1 [[user(locn1)]]; - float2 vColorTexCoord0 [[user(locn2)]]; - float4 vBaseColor [[user(locn3)]]; + float2 vColorTexCoord0 [[user(locn1)]]; + float4 vBaseColor [[user(locn2)]]; }; // Implementation of the GLSL mod() function, which is slightly different than Metal fmod() @@ -548,47 +545,42 @@ float4 composite(thread const float4& srcColor, thread const texture2d de return float4(((srcColor.xyz * (srcColor.w * (1.0 - destColor.w))) + (blendedRGB * (srcColor.w * destColor.w))) + (destColor.xyz * (1.0 - srcColor.w)), 1.0); } -void calculateColor(thread const int& ctrl, thread texture2d uMaskTexture0, thread const sampler uMaskTexture0Smplr, thread float3& vMaskTexCoord0, thread texture2d uMaskTexture1, thread const sampler uMaskTexture1Smplr, thread float3& vMaskTexCoord1, thread float4& vBaseColor, thread float2& vColorTexCoord0, thread texture2d uColorTexture0, thread const sampler uColorTexture0Smplr, thread texture2d uGammaLUT, thread const sampler uGammaLUTSmplr, thread float2 uColorTexture0Size, thread float4& gl_FragCoord, thread float2 uFramebufferSize, thread float4 uFilterParams0, thread float4 uFilterParams1, thread float4 uFilterParams2, thread texture2d uDestTexture, thread const sampler uDestTextureSmplr, thread float4& oFragColor) +void calculateColor(thread const int& ctrl, thread texture2d uMaskTexture0, thread const sampler uMaskTexture0Smplr, thread float3& vMaskTexCoord0, thread float4& vBaseColor, thread float2& vColorTexCoord0, thread texture2d uColorTexture0, thread const sampler uColorTexture0Smplr, thread texture2d uGammaLUT, thread const sampler uGammaLUTSmplr, thread float2 uColorTexture0Size, thread float4& gl_FragCoord, thread float2 uFramebufferSize, thread float4 uFilterParams0, thread float4 uFilterParams1, thread float4 uFilterParams2, thread texture2d uDestTexture, thread const sampler uDestTextureSmplr, thread float4& oFragColor) { int maskCtrl0 = (ctrl >> 0) & 3; - int maskCtrl1 = (ctrl >> 2) & 3; float maskAlpha = 1.0; float param = maskAlpha; float3 param_1 = vMaskTexCoord0; int param_2 = maskCtrl0; maskAlpha = sampleMask(param, uMaskTexture0, uMaskTexture0Smplr, param_1, param_2); - float param_3 = maskAlpha; - float3 param_4 = vMaskTexCoord1; - int param_5 = maskCtrl1; - maskAlpha = sampleMask(param_3, uMaskTexture1, uMaskTexture1Smplr, param_4, param_5); float4 color = vBaseColor; int color0Combine = (ctrl >> 6) & 3; if (color0Combine != 0) { int color0Filter = (ctrl >> 4) & 3; - float2 param_6 = vColorTexCoord0; - float2 param_7 = uColorTexture0Size; - float2 param_8 = gl_FragCoord.xy; - float2 param_9 = uFramebufferSize; - float4 param_10 = uFilterParams0; - float4 param_11 = uFilterParams1; - float4 param_12 = uFilterParams2; - int param_13 = color0Filter; - float4 color0 = filterColor(param_6, uColorTexture0, uColorTexture0Smplr, uGammaLUT, uGammaLUTSmplr, param_7, param_8, param_9, param_10, param_11, param_12, param_13); - float4 param_14 = color; - float4 param_15 = color0; - int param_16 = color0Combine; - color = combineColor0(param_14, param_15, param_16); + float2 param_3 = vColorTexCoord0; + float2 param_4 = uColorTexture0Size; + float2 param_5 = gl_FragCoord.xy; + float2 param_6 = uFramebufferSize; + float4 param_7 = uFilterParams0; + float4 param_8 = uFilterParams1; + float4 param_9 = uFilterParams2; + int param_10 = color0Filter; + float4 color0 = filterColor(param_3, uColorTexture0, uColorTexture0Smplr, uGammaLUT, uGammaLUTSmplr, param_4, param_5, param_6, param_7, param_8, param_9, param_10); + float4 param_11 = color; + float4 param_12 = color0; + int param_13 = color0Combine; + color = combineColor0(param_11, param_12, param_13); } color.w *= maskAlpha; int compositeOp = (ctrl >> 8) & 15; - float4 param_17 = color; - float2 param_18 = uFramebufferSize; - float2 param_19 = gl_FragCoord.xy; - int param_20 = compositeOp; - color = composite(param_17, uDestTexture, uDestTextureSmplr, param_18, param_19, param_20); - float3 _1337 = color.xyz * color.w; - color = float4(_1337.x, _1337.y, _1337.z, color.w); + float4 param_14 = color; + float2 param_15 = uFramebufferSize; + float2 param_16 = gl_FragCoord.xy; + int param_17 = compositeOp; + color = composite(param_14, uDestTexture, uDestTextureSmplr, param_15, param_16, param_17); + float3 _1324 = color.xyz * color.w; + color = float4(_1324.x, _1324.y, _1324.z, color.w); oFragColor = color; } @@ -596,7 +588,7 @@ fragment main0_out main0(main0_in in [[stage_in]], constant spvDescriptorSetBuff { main0_out out = {}; int param = (*spvDescriptorSet0.uCtrl); - calculateColor(param, spvDescriptorSet0.uMaskTexture0, spvDescriptorSet0.uMaskTexture0Smplr, in.vMaskTexCoord0, spvDescriptorSet0.uMaskTexture1, spvDescriptorSet0.uMaskTexture1Smplr, in.vMaskTexCoord1, in.vBaseColor, in.vColorTexCoord0, spvDescriptorSet0.uColorTexture0, spvDescriptorSet0.uColorTexture0Smplr, spvDescriptorSet0.uGammaLUT, spvDescriptorSet0.uGammaLUTSmplr, (*spvDescriptorSet0.uColorTexture0Size), gl_FragCoord, (*spvDescriptorSet0.uFramebufferSize), (*spvDescriptorSet0.uFilterParams0), (*spvDescriptorSet0.uFilterParams1), (*spvDescriptorSet0.uFilterParams2), spvDescriptorSet0.uDestTexture, spvDescriptorSet0.uDestTextureSmplr, out.oFragColor); + calculateColor(param, spvDescriptorSet0.uMaskTexture0, spvDescriptorSet0.uMaskTexture0Smplr, in.vMaskTexCoord0, in.vBaseColor, in.vColorTexCoord0, spvDescriptorSet0.uColorTexture0, spvDescriptorSet0.uColorTexture0Smplr, spvDescriptorSet0.uGammaLUT, spvDescriptorSet0.uGammaLUTSmplr, (*spvDescriptorSet0.uColorTexture0Size), gl_FragCoord, (*spvDescriptorSet0.uFramebufferSize), (*spvDescriptorSet0.uFilterParams0), (*spvDescriptorSet0.uFilterParams1), (*spvDescriptorSet0.uFilterParams2), spvDescriptorSet0.uDestTexture, spvDescriptorSet0.uDestTextureSmplr, out.oFragColor); return out; } diff --git a/resources/shaders/metal/tile.vs.metal b/resources/shaders/metal/tile.vs.metal index 402a929f..7ce7f360 100644 --- a/resources/shaders/metal/tile.vs.metal +++ b/resources/shaders/metal/tile.vs.metal @@ -16,9 +16,8 @@ struct spvDescriptorSetBuffer0 struct main0_out { float3 vMaskTexCoord0 [[user(locn0)]]; - float3 vMaskTexCoord1 [[user(locn1)]]; - float2 vColorTexCoord0 [[user(locn2)]]; - float4 vBaseColor [[user(locn3)]]; + float2 vColorTexCoord0 [[user(locn1)]]; + float4 vBaseColor [[user(locn2)]]; float4 gl_Position [[position]]; }; @@ -27,9 +26,8 @@ struct main0_in int2 aTileOffset [[attribute(0)]]; int2 aTileOrigin [[attribute(1)]]; uint2 aMaskTexCoord0 [[attribute(2)]]; - uint2 aMaskTexCoord1 [[attribute(3)]]; - int2 aMaskBackdrop [[attribute(4)]]; - int aColor [[attribute(5)]]; + int2 aMaskBackdrop [[attribute(3)]]; + int aColor [[attribute(4)]]; }; vertex main0_out main0(main0_in in [[stage_in]], constant spvDescriptorSetBuffer0& spvDescriptorSet0 [[buffer(0)]]) @@ -39,7 +37,6 @@ vertex main0_out main0(main0_in in [[stage_in]], constant spvDescriptorSetBuffer float2 tileOffset = float2(in.aTileOffset); float2 position = (tileOrigin + tileOffset) * (*spvDescriptorSet0.uTileSize); float2 maskTexCoord0 = (float2(in.aMaskTexCoord0) + tileOffset) / float2(256.0); - float2 maskTexCoord1 = (float2(in.aMaskTexCoord1) + tileOffset) / float2(256.0); float2 textureMetadataScale = float2(1.0) / float2((*spvDescriptorSet0.uTextureMetadataSize)); float2 metadataEntryCoord = float2(float((in.aColor % 128) * 4), float(in.aColor / 128)); float2 colorTexMatrix0Coord = (metadataEntryCoord + float2(0.5)) * textureMetadataScale; @@ -50,7 +47,6 @@ vertex main0_out main0(main0_in in [[stage_in]], constant spvDescriptorSetBuffer float4 baseColor = spvDescriptorSet0.uTextureMetadata.sample(spvDescriptorSet0.uTextureMetadataSmplr, baseColorCoord, level(0.0)); out.vColorTexCoord0 = (float2x2(float2(colorTexMatrix0.xy), float2(colorTexMatrix0.zw)) * position) + colorTexOffsets.xy; out.vMaskTexCoord0 = float3(maskTexCoord0, float(in.aMaskBackdrop.x)); - out.vMaskTexCoord1 = float3(maskTexCoord1, float(in.aMaskBackdrop.y)); out.vBaseColor = baseColor; out.gl_Position = (*spvDescriptorSet0.uTransform) * float4(position, 0.0, 1.0); return out; diff --git a/shaders/Makefile b/shaders/Makefile index 64535afe..f42582ad 100644 --- a/shaders/Makefile +++ b/shaders/Makefile @@ -19,6 +19,8 @@ SHADERS=\ stencil.vs.glsl \ tile.fs.glsl \ tile.vs.glsl \ + tile_clip.fs.glsl \ + tile_clip.vs.glsl \ tile_copy.fs.glsl \ tile_copy.vs.glsl \ $(EMPTY) diff --git a/shaders/tile.fs.glsl b/shaders/tile.fs.glsl index e795cb5c..12b2e5ea 100644 --- a/shaders/tile.fs.glsl +++ b/shaders/tile.fs.glsl @@ -73,14 +73,12 @@ precision highp sampler2D; #define COMBINER_CTRL_COMPOSITE_LUMINOSITY 0xf #define COMBINER_CTRL_MASK_0_SHIFT 0 -#define COMBINER_CTRL_MASK_1_SHIFT 2 #define COMBINER_CTRL_COLOR_FILTER_SHIFT 4 #define COMBINER_CTRL_COLOR_COMBINE_SHIFT 6 #define COMBINER_CTRL_COMPOSITE_SHIFT 8 uniform sampler2D uColorTexture0; uniform sampler2D uMaskTexture0; -uniform sampler2D uMaskTexture1; uniform sampler2D uDestTexture; uniform sampler2D uGammaLUT; uniform vec4 uFilterParams0; @@ -91,7 +89,6 @@ uniform vec2 uColorTexture0Size; uniform int uCtrl; in vec3 vMaskTexCoord0; -in vec3 vMaskTexCoord1; in vec2 vColorTexCoord0; in vec4 vBaseColor; @@ -560,10 +557,8 @@ float sampleMask(float maskAlpha, void calculateColor(int ctrl) { // Sample mask. int maskCtrl0 = (ctrl >> COMBINER_CTRL_MASK_0_SHIFT) & COMBINER_CTRL_MASK_MASK; - int maskCtrl1 = (ctrl >> COMBINER_CTRL_MASK_1_SHIFT) & COMBINER_CTRL_MASK_MASK; float maskAlpha = 1.0; maskAlpha = sampleMask(maskAlpha, uMaskTexture0, vMaskTexCoord0, maskCtrl0); - maskAlpha = sampleMask(maskAlpha, uMaskTexture1, vMaskTexCoord1, maskCtrl1); // Sample color. vec4 color = vBaseColor; diff --git a/shaders/tile.vs.glsl b/shaders/tile.vs.glsl index acbda0f0..769b8ac1 100644 --- a/shaders/tile.vs.glsl +++ b/shaders/tile.vs.glsl @@ -21,12 +21,10 @@ uniform ivec2 uTextureMetadataSize; in ivec2 aTileOffset; in ivec2 aTileOrigin; in uvec2 aMaskTexCoord0; -in uvec2 aMaskTexCoord1; in ivec2 aMaskBackdrop; in int aColor; out vec3 vMaskTexCoord0; -out vec3 vMaskTexCoord1; out vec2 vColorTexCoord0; out vec4 vBaseColor; @@ -35,7 +33,6 @@ void main() { vec2 position = (tileOrigin + tileOffset) * uTileSize; vec2 maskTexCoord0 = (vec2(aMaskTexCoord0) + tileOffset) / 256.0; - vec2 maskTexCoord1 = (vec2(aMaskTexCoord1) + tileOffset) / 256.0; vec2 textureMetadataScale = vec2(1.0) / vec2(uTextureMetadataSize); vec2 metadataEntryCoord = vec2(aColor % 128 * 4, aColor / 128); @@ -48,7 +45,6 @@ void main() { vColorTexCoord0 = mat2(colorTexMatrix0) * position + colorTexOffsets.xy; vMaskTexCoord0 = vec3(maskTexCoord0, float(aMaskBackdrop.x)); - vMaskTexCoord1 = vec3(maskTexCoord1, float(aMaskBackdrop.y)); vBaseColor = baseColor; gl_Position = uTransform * vec4(position, 0.0, 1.0); } diff --git a/shaders/tile_clip.fs.glsl b/shaders/tile_clip.fs.glsl new file mode 100644 index 00000000..3d6d02f9 --- /dev/null +++ b/shaders/tile_clip.fs.glsl @@ -0,0 +1,26 @@ +#version 330 + +// pathfinder/shaders/tile_clip.fs.glsl +// +// Copyright © 2020 The Pathfinder Project Developers. +// +// Licensed under the Apache License, Version 2.0 or the MIT license +// , at your +// option. This file may not be copied, modified, or distributed +// except according to those terms. + +precision highp float; +precision highp sampler2D; + +uniform sampler2D uSrc; + +in vec2 vTexCoord; +in float vBackdrop; + +out vec4 oFragColor; + +void main() { + float alpha = clamp(abs(texture(uSrc, vTexCoord).r + vBackdrop), 0.0, 1.0); + oFragColor = vec4(alpha, 0.0, 0.0, 1.0); +} diff --git a/shaders/tile_clip.vs.glsl b/shaders/tile_clip.vs.glsl new file mode 100644 index 00000000..38a8c450 --- /dev/null +++ b/shaders/tile_clip.vs.glsl @@ -0,0 +1,30 @@ +#version 330 + +// pathfinder/shaders/tile_clip.vs.glsl +// +// Copyright © 2020 The Pathfinder Project Developers. +// +// Licensed under the Apache License, Version 2.0 or the MIT license +// , at your +// option. This file may not be copied, modified, or distributed +// except according to those terms. + +precision highp float; +precision highp sampler2D; + +in ivec2 aTileOffset; +in ivec2 aDestTileOrigin; +in ivec2 aSrcTileOrigin; +in int aSrcBackdrop; + +out vec2 vTexCoord; +out float vBackdrop; + +void main() { + vec2 destPosition = vec2(aDestTileOrigin + aTileOffset) / vec2(256.0); + vec2 srcPosition = vec2(aSrcTileOrigin + aTileOffset) / vec2(256.0); + vTexCoord = srcPosition; + vBackdrop = float(aSrcBackdrop); + gl_Position = vec4(mix(vec2(-1.0), vec2(1.0), destPosition), 0.0, 1.0); +}