Repository: ruffle-rs/ruffle (language: rust)

Issue:
Cosmo Cabs Crash
https://www.newgrounds.com/portal/view/427349/format/flash?emulate=flash

After getting through the instructions screen, Cosmo Cabs crashes with this error:

# Error Info
Error name: Error
Error message: panicked at /home/runner/.cargo/registry/src/index.crates.io-6f17d22bba15001f/wgpu-hal-22.0.0/src/gles/device.rs:779:52:
called `Result::unwrap()` on an `Err` value: "Unable to create texture object"
Error stack:
```
Error: panicked at /home/runner/.cargo/registry/src/index.crates.io-6f17d22bba15001f/wgpu-hal-22.0.0/src/gles/device.rs:779:52:
called `Result::unwrap()` on an `Err` value: "Unable to create texture object"
    at n.wbg.__wbg_new_796382978dfd4fb0 (https://uploads.ungrounded.net/ruffle_wrapper/lib/core.ruffle.80738d45489f7f5f6247.js:1:84118)
    at ruffle_web.wasm.js_sys::Error::new::h6cc27b63b536b514 (wasm://wasm/ruffle_web.wasm-0328e5de:wasm-function[12770]:0x9a2161)
    at ruffle_web.wasm.core::ops::function::FnOnce::call_once{{vtable.shim}}::h97b75757bc3b995e (wasm://wasm/ruffle_web.wasm-0328e5de:wasm-function[9755]:0x932552)
    at ruffle_web.wasm.std::panicking::rust_panic_with_hook::he5c089ac7305193e (wasm://wasm/ruffle_web.wasm-0328e5de:wasm-function[6457]:0x832085)
    at ruffle_web.wasm.core::panicking::panic_fmt::hdc8d2d914c0710e4 (wasm://wasm/ruffle_web.wasm-0328e5de:wasm-function[9480]:0x923112)
    at ruffle_web.wasm.core::result::unwrap_failed::hdbb420d9dacf5a62 (wasm://wasm/ruffle_web.wasm-0328e5de:wasm-function[7966]:0x8afbaf)
    at ruffle_web.wasm.<T as wgpu::context::DynContext>::device_create_texture::h49957921db6a1784 (wasm://wasm/ruffle_web.wasm-0328e5de:wasm-function[7457]:0x88599a)
    at ruffle_web.wasm.wgpu::Device::create_texture::h9ff723e1379eae2e (wasm://wasm/ruffle_web.wasm-0328e5de:wasm-function[5788]:0x7fd170)
    at ruffle_web.wasm.<ruffle_render_wgpu::backend::WgpuRenderBackend<T> as ruffle_render::backend::RenderBackend>::register_bitmap::hfdf1871484edee70 (wasm://wasm/ruffle_web.wasm-0328e5de:wasm-function[2108]:0x529858)
    at ruffle_web.wasm.ruffle_core::bitmap::bitmap_data::BitmapData::bitmap_handle::h60f13afef21af52b (wasm://wasm/ruffle_web.wasm-0328e5de:wasm-function[2156]:0x538de0)
```
AVM2 stack:
```
    at flash.display::BitmapData/flash::display::BitmapData::draw()
    at net.superflashbros.utils::FullHitTest$/objectHitMask()
    at net.superflashbros.game::Terrain/fullHitTest()
    at net.superflashbros.game::Terrain/hitTest()
    at net.superflashbros.game::World/render()
    at flash.events::EventDispatcher/flash::events::EventDispatcher::dispatchEvent()
    at flash.utils::Timer/onUpdate()
```

# Player Info
Allows script access: false
Renderer: wgpu
Adapter Backend: Gl
Adapter Name: "ANGLE (NVIDIA, NVIDIA GeForce RTX 3080 (0x00002206) Direct3D11 vs_5_0 ps_5_0, D3D11)"
Adapter Device Type: Other
Adapter Driver Name: ""
Adapter Driver Info: "WebGL 2.0 (OpenGL ES 3.0 Chromium)"
Enabled features: Features(TEXTURE_COMPRESSION_BC | FLOAT32_FILTERABLE | TEXTURE_ADAPTER_SPECIFIC_FORMAT_FEATURES)
Available features: Features(DEPTH32FLOAT_STENCIL8 | PUSH_CONSTANTS | CLEAR_TEXTURE | MULTIVIEW)
Current limits: Limits { max_texture_dimension_1d: 16384, max_texture_dimension_2d: 16384, max_texture_dimension_3d: 2048, max_texture_array_layers: 256, max_bind_groups: 4, max_bindings_per_bind_group: 1000, max_dynamic_uniform_buffers_per_pipeline_layout: 8, max_dynamic_storage_buffers_per_pipeline_layout: 0, max_sampled_textures_per_shader_stage: 16, max_samplers_per_shader_stage: 16, max_storage_buffers_per_shader_stage: 0, max_storage_textures_per_shader_stage: 0, max_uniform_buffers_per_shader_stage: 11, max_uniform_buffer_binding_size: 65536, max_storage_buffer_binding_size: 0, max_vertex_buffers: 8, max_buffer_size: 268435456, max_vertex_attributes: 16, max_vertex_buffer_array_stride: 255, min_uniform_buffer_offset_alignment: 256, min_storage_buffer_offset_alignment: 256, max_inter_stage_shader_components: 120, max_color_attachments: 8, max_color_attachment_bytes_per_sample: 32, max_compute_workgroup_storage_size: 0, max_compute_invocations_per_workgroup: 0, max_compute_workgroup_size_x: 0, max_compute_workgroup_size_y: 0, max_compute_workgroup_size_z: 0, max_compute_workgroups_per_dimension: 0, min_subgroup_size: 0, max_subgroup_size: 0, max_push_constant_size: 0, max_non_sampler_bindings: 1000000 }
Surface quality: high
Surface samples: 4
Surface size: (550, 400, 1)

# Page Info
Page URL: https://uploads.ungrounded.net/ruffle_wrapper/ruffleembed.html?v=1.0.67&browsermode=default&props=%7B%22ruffle%22%3A%22%2Fruffle_wrapper%2Flib%2Fruffle.js%3F1729224944%22%2C%22public_path%22%3A%22%2Fruffle_wrapper%2Flib%2F%22%2C%22swf%22%3A%22https%3A%2F%2Fuploads.ungrounded.net%2F427000%2F427349_Cosmo_Cabs_v03.swf%3F1204074437%22%2C%22vars%22%3A%7B%22NewgroundsAPI_PublisherID%22%3A1%2C%22NewgroundsAPI_SandboxID%22%3A%2267130bb9df224%22%2C%22NewgroundsAPI_SessionID%22%3A%22G7pS484BC4F1AXLK3V2s356a192b7913b04c54574d18c28d46e6395428abqKCh%22%2C%22NewgroundsAPI_UserName%22%3A%22TomFulp%22%2C%22NewgroundsAPI_UserID%22%3A1%2C%22ng_username%22%3A%22TomFulp%22%7D%2C%22width%22%3A550%2C%22height%22%3A400%2C%22icon%22%3A%22https%3A%2F%2Fpicon.ngfiles.com%2F427000%2Fflash_427349.gif%3Ff1276814763%22%2C%22warnOnUnsupportedContent%22%3Atrue%7D
SWF URL: https://uploads.ungrounded.net/427000/427349_Cosmo_Cabs_v03.swf?1204074437

# Browser Info
User Agent: Mozilla/5.0 (Windows NT 10.0; Win64; x64) AppleWebKit/537.36 (KHTML, like Gecko) Chrome/129.0.0.0 Safari/537.36
Platform: Win32
Has touch support: false

# Ruffle Info
Version: 0.1.0
Name: nightly 2024-10-18
Channel: nightly
Built: 2024-10-18T00:06:13.525Z
Commit: 5faa313a70d6ad47af18ad4b2f6517696118ef07
Is extension: false

# Metadata
width: 550
height: 400
frameRate: 15
numFrames: 8
swfVersion: 9
backgroundColor: #16181D
isActionScript3: true
uncompressedLength: 1401295



Discussion / hints:
Yeah, it's us being super inefficient.
The single `bitmapdata.draw` tries to draw a displayobject with 1000+ children with `BlendType.OVERLAY`, each one trying to allocate a 2314x1328 texture because of our current blending implementation; which adds up to ~20GB of VRAM usage.
We do have rough ideas for improving blending (we definitely shouldn't need a full 2314x1328 texture for each separate blend here, making it smaller would reduce memory use by ~1000x), but they've not figured out fully how to do them yet.

Fix patch:
diff --git a/Cargo.lock b/Cargo.lock
index a69ea3495e42b..71f13f86f57cf 100644
--- a/Cargo.lock
+++ b/Cargo.lock
@@ -2744,7 +2744,7 @@ source = "registry+https://github.com/rust-lang/crates.io-index"
 checksum = "4979f22fdb869068da03c9f7528f8297c6fd2606bc3a4affe42e6a823fdb8da4"
 dependencies = [
  "cfg-if",
- "windows-targets 0.52.6",
+ "windows-targets 0.48.5",
 ]
 
 [[package]]
@@ -5538,9 +5538,9 @@ dependencies = [
 
 [[package]]
 name = "tracy-client"
-version = "0.17.1"
+version = "0.17.4"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "63de1e1d4115534008d8fd5788b39324d6f58fc707849090533828619351d855"
+checksum = "746b078c6a09ebfd5594609049e07116735c304671eaab06ce749854d23435bc"
 dependencies = [
  "loom",
  "once_cell",
@@ -5550,11 +5550,12 @@ dependencies = [
 
 [[package]]
 name = "tracy-client-sys"
-version = "0.23.0"
+version = "0.24.2"
 source = "registry+https://github.com/rust-lang/crates.io-index"
-checksum = "98b98232a2447ce0a58f9a0bfb5f5e39647b5c597c994b63945fcccd1306fafb"
+checksum = "3637e734239e12ab152cd269302500bd063f37624ee210cd04b4936ed671f3b1"
 dependencies = [
  "cc",
+ "windows-targets 0.48.5",
 ]
 
 [[package]]
diff --git a/render/wgpu/src/buffer_builder.rs b/render/wgpu/src/buffer_builder.rs
index d042164ee19c4..82b49db2d1ffa 100644
--- a/render/wgpu/src/buffer_builder.rs
+++ b/render/wgpu/src/buffer_builder.rs
@@ -3,6 +3,7 @@ use std::ops::Range;
 use wgpu::util::DeviceExt;
 use wgpu::BufferAddress;
 
+#[derive(Debug)]
 pub struct BufferBuilder {
     inner: Vec<u8>,
     align_mask: usize,
diff --git a/render/wgpu/src/surface.rs b/render/wgpu/src/surface.rs
index ac093661d2b11..e02ed43cee918 100644
--- a/render/wgpu/src/surface.rs
+++ b/render/wgpu/src/surface.rs
@@ -2,27 +2,26 @@ mod commands;
 pub mod target;
 
 use crate::backend::RenderTargetMode;
-use crate::blend::ComplexBlend;
+use crate::blend::{BlendType, ComplexBlend};
 use crate::buffer_pool::TexturePool;
 use crate::dynamic_transforms::DynamicTransforms;
 use crate::filters::FilterSource;
 use crate::mesh::Mesh;
 use crate::pixel_bender::{run_pixelbender_shader_impl, ShaderMode};
+pub use crate::surface::commands::LayerRef;
 use crate::surface::commands::{chunk_blends, Chunk, CommandRenderer};
+use crate::utils::run_copy_pipeline;
 use crate::utils::{remove_srgb, supported_sample_count};
-use crate::{Descriptors, MaskState, Pipelines};
-use ruffle_render::commands::CommandList;
+use crate::{Descriptors, MaskState, Pipelines, Transforms};
+use ruffle_render::commands::{CommandList, RenderBlendMode};
+use ruffle_render::matrix::Matrix;
 use ruffle_render::pixel_bender::{ImageInputTexture, PixelBenderShaderArgument};
 use ruffle_render::quality::StageQuality;
 use std::sync::Arc;
+use swf::BlendMode;
 use target::CommandTarget;
 use tracing::instrument;
-
-use crate::utils::run_copy_pipeline;
-
-pub use crate::surface::commands::LayerRef;
-
-use self::commands::ChunkBlendMode;
+use wgpu::BufferAddress;
 
 #[derive(Debug)]
 pub struct Surface {
@@ -131,22 +130,7 @@ impl Surface {
 
         let mut num_masks = 0;
         let mut mask_state = MaskState::NoMask;
-        let chunks = chunk_blends(
-            commands,
-            descriptors,
-            staging_belt,
-            dynamic_transforms,
-            draw_encoder,
-            meshes,
-            self.quality,
-            target.width(),
-            target.height(),
-            match nearest_layer {
-                LayerRef::Current => LayerRef::Parent(&target),
-                layer => layer,
-            },
-            texture_pool,
-        );
+        let chunks = chunk_blends(commands, descriptors, dynamic_transforms);
 
         for chunk in chunks {
             match chunk {
@@ -194,44 +178,18 @@ impl Surface {
                     num_masks = renderer.num_masks();
                     mask_state = renderer.mask_state();
                 }
-                Chunk::Blend(texture, ChunkBlendMode::Shader(shader), needs_stencil) => {
-                    assert!(
-                        !needs_stencil,
-                        "Shader blend should not need stencil buffer"
-                    );
-                    let parent_blend_buffer =
-                        target.update_blend_buffer(descriptors, texture_pool, draw_encoder);
-                    run_pixelbender_shader_impl(
+                Chunk::Blend(chunk, needs_stencil, blend_mode) => {
+                    let is_layer = matches!(blend_mode, RenderBlendMode::Builtin(BlendMode::Layer));
+                    let blend_type = BlendType::from(blend_mode);
+                    let mut child = Surface::new(
                         descriptors,
-                        shader,
-                        ShaderMode::Filter,
-                        &[
-                            PixelBenderShaderArgument::ImageInput {
-                                index: 0,
-                                channels: 0xFF,
-                                name: "background".to_string(),
-                                texture: Some(ImageInputTexture::TextureRef(
-                                    parent_blend_buffer.texture(),
-                                )),
-                            },
-                            PixelBenderShaderArgument::ImageInput {
-                                index: 1,
-                                channels: 0xff,
-                                name: "foreground".to_string(),
-                                texture: Some(ImageInputTexture::TextureRef(texture.texture())),
-                            },
-                        ],
-                        parent_blend_buffer.texture(),
-                        draw_encoder,
-                        target.color_attachments(),
-                        target.sample_count(),
-                        &FilterSource::for_entire_texture(texture.texture()),
-                    )
-                    .expect("Failed to run PixelBender blend mode");
-                }
-                Chunk::Blend(texture, ChunkBlendMode::Complex(blend_mode), needs_stencil) => {
-                    let parent = match blend_mode {
-                        ComplexBlend::Alpha | ComplexBlend::Erase => {
+                        self.quality,
+                        target.width(),
+                        target.height(),
+                        wgpu::TextureFormat::Rgba8Unorm,
+                    );
+                    let parent = match blend_type {
+                        BlendType::Complex(ComplexBlend::Alpha | ComplexBlend::Erase) => {
                             match nearest_layer {
                                 LayerRef::None => {
                                     // An Alpha or Erase with no Layer above it should be ignored
@@ -243,102 +201,252 @@ impl Surface {
                         }
                         _ => &target,
                     };
-
                     let parent_blend_buffer =
                         parent.update_blend_buffer(descriptors, texture_pool, draw_encoder);
 
-                    let blend_bind_group =
-                        descriptors
-                            .device
-                            .create_bind_group(&wgpu::BindGroupDescriptor {
-                                label: create_debug_label!(
-                                    "Complex blend binds {:?} {}",
-                                    blend_mode,
-                                    if needs_stencil {
-                                        "(with stencil)"
-                                    } else {
-                                        "(Stencilless)"
-                                    }
+                    let child_target = child.draw_commands(
+                        RenderTargetMode::FreshWithColor(blend_type.default_color()),
+                        descriptors,
+                        meshes,
+                        chunk,
+                        staging_belt,
+                        dynamic_transforms,
+                        draw_encoder,
+                        match (is_layer, nearest_layer) {
+                            (true, _) => LayerRef::Current,
+                            (false, LayerRef::Current) => LayerRef::Parent(&target),
+                            (false, layer) => layer,
+                        },
+                        texture_pool,
+                    );
+                    child_target.ensure_cleared(draw_encoder);
+                    let child_texture = child_target.take_color_texture();
+
+                    match blend_type {
+                        BlendType::Trivial(blend_mode) => {
+                            let matrix =
+                                Matrix::scale(parent.width() as f32, parent.height() as f32);
+                            let transform = [Transforms {
+                                world_matrix: [
+                                    [matrix.a, matrix.b, 0.0, 0.0],
+                                    [matrix.c, matrix.d, 0.0, 0.0],
+                                    [0.0, 0.0, 1.0, 0.0],
+                                    [
+                                        matrix.tx.to_pixels() as f32,
+                                        matrix.ty.to_pixels() as f32,
+                                        0.0,
+                                        1.0,
+                                    ],
+                                ],
+                                mult_color: [1.0, 1.0, 1.0, 1.0],
+                                add_color: [0.0, 0.0, 0.0, 0.0],
+                            }];
+                            let transform_bytes: &[u8] = bytemuck::cast_slice(&transform);
+                            staging_belt
+                                .write_buffer(
+                                    draw_encoder,
+                                    &dynamic_transforms.buffer,
+                                    BufferAddress::default(),
+                                    wgpu::BufferSize::new(transform_bytes.len() as u64).unwrap(),
+                                    &descriptors.device,
                                 )
-                                .as_deref(),
-                                layout: &descriptors.bind_layouts.blend,
-                                entries: &[
-                                    wgpu::BindGroupEntry {
-                                        binding: 0,
-                                        resource: wgpu::BindingResource::TextureView(
-                                            parent_blend_buffer.view(),
-                                        ),
-                                    },
-                                    wgpu::BindGroupEntry {
-                                        binding: 1,
-                                        resource: wgpu::BindingResource::TextureView(
-                                            texture.view(),
-                                        ),
+                                .copy_from_slice(transform_bytes);
+                            let mut render_pass =
+                                draw_encoder.begin_render_pass(&wgpu::RenderPassDescriptor {
+                                    label: create_debug_label!(
+                                        "Apply trivial blend {blend_mode:?} {}",
+                                        if needs_stencil {
+                                            "(with stencil)"
+                                        } else {
+                                            "(Stencilless)"
+                                        }
+                                    )
+                                    .as_deref(),
+                                    color_attachments: &[target.color_attachments()],
+                                    depth_stencil_attachment: if needs_stencil {
+                                        target.stencil_attachment(descriptors, texture_pool)
+                                    } else {
+                                        None
                                     },
-                                    wgpu::BindGroupEntry {
-                                        binding: 2,
-                                        resource: wgpu::BindingResource::Sampler(
-                                            descriptors.bitmap_samplers.get_sampler(false, false),
-                                        ),
+                                    ..Default::default()
+                                });
+                            render_pass.set_bind_group(0, target.globals().bind_group(), &[]);
+                            let mut renderer = CommandRenderer::new(
+                                &self.pipelines,
+                                descriptors,
+                                dynamic_transforms,
+                                render_pass,
+                                num_masks,
+                                mask_state,
+                                needs_stencil,
+                            );
+                            let bind_group =
+                                descriptors
+                                    .device
+                                    .create_bind_group(&wgpu::BindGroupDescriptor {
+                                        layout: &descriptors.bind_layouts.bitmap,
+                                        entries: &[
+                                            wgpu::BindGroupEntry {
+                                                binding: 0,
+                                                resource: descriptors
+                                                    .quad
+                                                    .texture_transforms
+                                                    .as_entire_binding(),
+                                            },
+                                            wgpu::BindGroupEntry {
+                                                binding: 1,
+                                                resource: wgpu::BindingResource::TextureView(
+                                                    child_texture.view(),
+                                                ),
+                                            },
+                                            wgpu::BindGroupEntry {
+                                                binding: 2,
+                                                resource: wgpu::BindingResource::Sampler(
+                                                    descriptors
+                                                        .bitmap_samplers
+                                                        .get_sampler(false, false),
+                                                ),
+                                            },
+                                        ],
+                                        label: None,
+                                    });
+                            renderer.render_texture(0, &bind_group, blend_mode);
+                        }
+                        BlendType::Complex(blend_mode) => {
+                            let blend_bind_group =
+                                descriptors
+                                    .device
+                                    .create_bind_group(&wgpu::BindGroupDescriptor {
+                                        label: create_debug_label!(
+                                            "Complex blend binds {:?} {}",
+                                            blend_mode,
+                                            if needs_stencil {
+                                                "(with stencil)"
+                                            } else {
+                                                "(Stencilless)"
+                                            }
+                                        )
+                                        .as_deref(),
+                                        layout: &descriptors.bind_layouts.blend,
+                                        entries: &[
+                                            wgpu::BindGroupEntry {
+                                                binding: 0,
+                                                resource: wgpu::BindingResource::TextureView(
+                                                    parent_blend_buffer.view(),
+                                                ),
+                                            },
+                                            wgpu::BindGroupEntry {
+                                                binding: 1,
+                                                resource: wgpu::BindingResource::TextureView(
+                                                    child_texture.view(),
+                                                ),
+                                            },
+                                            wgpu::BindGroupEntry {
+                                                binding: 2,
+                                                resource: wgpu::BindingResource::Sampler(
+                                                    descriptors
+                                                        .bitmap_samplers
+                                                        .get_sampler(false, false),
+                                                ),
+                                            },
+                                        ],
+                                    });
+
+                            let mut render_pass =
+                                draw_encoder.begin_render_pass(&wgpu::RenderPassDescriptor {
+                                    label: create_debug_label!(
+                                        "Apply complex blend {:?} {}",
+                                        blend_mode,
+                                        if needs_stencil {
+                                            "(with stencil)"
+                                        } else {
+                                            "(Stencilless)"
+                                        }
+                                    )
+                                    .as_deref(),
+                                    color_attachments: &[target.color_attachments()],
+                                    depth_stencil_attachment: if needs_stencil {
+                                        target.stencil_attachment(descriptors, texture_pool)
+                                    } else {
+                                        None
                                     },
-                                ],
-                            });
+                                    ..Default::default()
+                                });
+                            render_pass.set_bind_group(0, target.globals().bind_group(), &[]);
 
-                    let mut render_pass =
-                        draw_encoder.begin_render_pass(&wgpu::RenderPassDescriptor {
-                            label: create_debug_label!(
-                                "Complex blend {:?} {}",
-                                blend_mode,
-                                if needs_stencil {
-                                    "(with stencil)"
-                                } else {
-                                    "(Stencilless)"
+                            if needs_stencil {
+                                match mask_state {
+                                    MaskState::NoMask => {}
+                                    MaskState::DrawMaskStencil => {
+                                        render_pass.set_stencil_reference(num_masks - 1);
+                                    }
+                                    MaskState::DrawMaskedContent => {
+                                        render_pass.set_stencil_reference(num_masks);
+                                    }
+                                    MaskState::ClearMaskStencil => {
+                                        render_pass.set_stencil_reference(num_masks);
+                                    }
                                 }
-                            )
-                            .as_deref(),
-                            color_attachments: &[target.color_attachments()],
-                            depth_stencil_attachment: if needs_stencil {
-                                target.stencil_attachment(descriptors, texture_pool)
+                                render_pass.set_pipeline(
+                                    self.pipelines.complex_blends[blend_mode]
+                                        .pipeline_for(mask_state),
+                                );
                             } else {
-                                None
-                            },
-                            ..Default::default()
-                        });
-                    render_pass.set_bind_group(0, target.globals().bind_group(), &[]);
-
-                    if needs_stencil {
-                        match mask_state {
-                            MaskState::NoMask => {}
-                            MaskState::DrawMaskStencil => {
-                                render_pass.set_stencil_reference(num_masks - 1);
+                                render_pass.set_pipeline(
+                                    self.pipelines.complex_blends[blend_mode]
+                                        .stencilless_pipeline(),
+                                );
                             }
-                            MaskState::DrawMaskedContent => {
-                                render_pass.set_stencil_reference(num_masks);
-                            }
-                            MaskState::ClearMaskStencil => {
-                                render_pass.set_stencil_reference(num_masks);
-                            }
-                        }
-                        render_pass.set_pipeline(
-                            self.pipelines.complex_blends[blend_mode].pipeline_for(mask_state),
-                        );
-                    } else {
-                        render_pass.set_pipeline(
-                            self.pipelines.complex_blends[blend_mode].stencilless_pipeline(),
-                        );
-                    }
 
-                    render_pass.set_bind_group(1, target.whole_frame_bind_group(descriptors), &[0]);
-                    render_pass.set_bind_group(2, &blend_bind_group, &[]);
+                            render_pass.set_bind_group(
+                                1,
+                                target.whole_frame_bind_group(descriptors),
+                                &[0],
+                            );
+                            render_pass.set_bind_group(2, &blend_bind_group, &[]);
 
-                    render_pass.set_vertex_buffer(0, descriptors.quad.vertices_pos.slice(..));
-                    render_pass.set_index_buffer(
-                        descriptors.quad.indices.slice(..),
-                        wgpu::IndexFormat::Uint32,
-                    );
+                            render_pass
+                                .set_vertex_buffer(0, descriptors.quad.vertices_pos.slice(..));
+                            render_pass.set_index_buffer(
+                                descriptors.quad.indices.slice(..),
+                                wgpu::IndexFormat::Uint32,
+                            );
 
-                    render_pass.draw_indexed(0..6, 0, 0..1);
-                    drop(render_pass);
+                            render_pass.draw_indexed(0..6, 0, 0..1);
+                        }
+                        BlendType::Shader(shader) => {
+                            run_pixelbender_shader_impl(
+                                descriptors,
+                                shader,
+                                ShaderMode::Filter,
+                                &[
+                                    PixelBenderShaderArgument::ImageInput {
+                                        index: 0,
+                                        channels: 0xFF,
+                                        name: "background".to_string(),
+                                        texture: Some(ImageInputTexture::TextureRef(
+                                            parent_blend_buffer.texture(),
+                                        )),
+                                    },
+                                    PixelBenderShaderArgument::ImageInput {
+                                        index: 1,
+                                        channels: 0xff,
+                                        name: "foreground".to_string(),
+                                        texture: Some(ImageInputTexture::TextureRef(
+                                            child_texture.texture(),
+                                        )),
+                                    },
+                                ],
+                                parent_blend_buffer.texture(),
+                                draw_encoder,
+                                target.color_attachments(),
+                                target.sample_count(),
+                                &FilterSource::for_entire_texture(child_texture.texture()),
+                            )
+                            .expect("Failed to run PixelBender blend mode");
+                        }
+                    }
+                    drop(child_texture);
                 }
             }
         }
diff --git a/render/wgpu/src/surface/commands.rs b/render/wgpu/src/surface/commands.rs
index e708c719d9049..c126cb47e283f 100644
--- a/render/wgpu/src/surface/commands.rs
+++ b/render/wgpu/src/surface/commands.rs
@@ -1,27 +1,19 @@
-use crate::backend::RenderTargetMode;
 use crate::blend::TrivialBlend;
-use crate::blend::{BlendType, ComplexBlend};
 use crate::buffer_builder::BufferBuilder;
-use crate::buffer_pool::TexturePool;
 use crate::dynamic_transforms::DynamicTransforms;
-use crate::mesh::{as_mesh, DrawType, Mesh};
+use crate::mesh::{as_mesh, DrawType};
 use crate::surface::target::CommandTarget;
-use crate::surface::Surface;
 use crate::{as_texture, Descriptors, MaskState, Pipelines, Transforms};
 use ruffle_render::backend::ShapeHandle;
 use ruffle_render::bitmap::{BitmapHandle, PixelSnapping};
 use ruffle_render::commands::{CommandHandler, CommandList, RenderBlendMode};
 use ruffle_render::lines::{emulate_line, emulate_line_rect};
 use ruffle_render::matrix::Matrix;
-use ruffle_render::pixel_bender::PixelBenderShaderHandle;
-use ruffle_render::quality::StageQuality;
 use ruffle_render::transform::Transform;
 use std::mem;
-use swf::{BlendMode, Color, ColorTransform, Twips};
+use swf::{Color, ColorTransform, Twips};
 use wgpu::Backend;
 
-use super::target::PoolOrArcTexture;
-
 pub struct CommandRenderer<'pass, 'frame: 'pass, 'global: 'frame> {
     pipelines: &'frame Pipelines,
     descriptors: &'global Descriptors,
@@ -84,12 +76,6 @@ impl<'pass, 'frame: 'pass, 'global: 'frame> CommandRenderer<'pass, 'frame, 'glob
                 *blend_mode,
                 *render_stage3d,
             ),
-            DrawCommand::RenderTexture {
-                _texture,
-                binds,
-                transform_buffer,
-                blend_mode,
-            } => self.render_texture(*transform_buffer, binds, *blend_mode),
             DrawCommand::RenderShape {
                 shape,
                 transform_buffer,
@@ -387,15 +373,10 @@ impl<'pass, 'frame: 'pass, 'global: 'frame> CommandRenderer<'pass, 'frame, 'glob
     }
 }
 
+#[derive(Debug)]
 pub enum Chunk {
     Draw(Vec<DrawCommand>, bool, BufferBuilder),
-    Blend(PoolOrArcTexture, ChunkBlendMode, bool),
-}
-
-#[derive(Debug)]
-pub enum ChunkBlendMode {
-    Complex(ComplexBlend),
-    Shader(PixelBenderShaderHandle),
+    Blend(CommandList, bool, RenderBlendMode),
 }
 
 #[derive(Debug)]
@@ -407,12 +388,6 @@ pub enum DrawCommand {
         blend_mode: TrivialBlend,
         render_stage3d: bool,
     },
-    RenderTexture {
-        _texture: PoolOrArcTexture,
-        binds: wgpu::BindGroup,
-        transform_buffer: wgpu::DynamicOffset,
-        blend_mode: TrivialBlend,
-    },
     RenderShape {
         shape: ShapeHandle,
         transform_buffer: wgpu::DynamicOffset,
@@ -439,48 +414,20 @@ pub enum LayerRef<'a> {
     Parent(&'a CommandTarget),
 }
 
-/// Replaces every blend with a RenderBitmap, with the subcommands rendered out to a temporary texture
-/// Every complex blend will be its own item, but every other draw will be chunked together
+/// Separates the command list into things that can be batched into the same render pass.
+/// Every blend will be its own item, but every other draw will be chunked together
 #[allow(clippy::too_many_arguments)]
 pub fn chunk_blends<'a>(
     commands: CommandList,
     descriptors: &'a Descriptors,
-    staging_belt: &'a mut wgpu::util::StagingBelt,
     dynamic_transforms: &'a DynamicTransforms,
-    draw_encoder: &mut wgpu::CommandEncoder,
-    meshes: &'a Vec<Mesh>,
-    quality: StageQuality,
-    width: u32,
-    height: u32,
-    nearest_layer: LayerRef,
-    texture_pool: &mut TexturePool,
 ) -> Vec<Chunk> {
-    WgpuCommandHandler::new(
-        descriptors,
-        staging_belt,
-        dynamic_transforms,
-        draw_encoder,
-        meshes,
-        quality,
-        width,
-        height,
-        nearest_layer,
-        texture_pool,
-    )
-    .chunk_blends(commands)
+    WgpuCommandHandler::new(descriptors, dynamic_transforms).chunk_blends(commands)
 }
 
 struct WgpuCommandHandler<'a> {
     descriptors: &'a Descriptors,
-    quality: StageQuality,
-    width: u32,
-    height: u32,
-    nearest_layer: LayerRef<'a>,
-    meshes: &'a Vec<Mesh>,
-    staging_belt: &'a mut wgpu::util::StagingBelt,
     dynamic_transforms: &'a DynamicTransforms,
-    draw_encoder: &'a mut wgpu::CommandEncoder,
-    texture_pool: &'a mut TexturePool,
     emulate_lines: bool,
 
     result: Vec<Chunk>,
@@ -492,18 +439,7 @@ struct WgpuCommandHandler<'a> {
 
 impl<'a> WgpuCommandHandler<'a> {
     #[allow(clippy::too_many_arguments)]
-    fn new(
-        descriptors: &'a Descriptors,
-        staging_belt: &'a mut wgpu::util::StagingBelt,
-        dynamic_transforms: &'a DynamicTransforms,
-        draw_encoder: &'a mut wgpu::CommandEncoder,
-        meshes: &'a Vec<Mesh>,
-        quality: StageQuality,
-        width: u32,
-        height: u32,
-        nearest_layer: LayerRef<'a>,
-        texture_pool: &'a mut TexturePool,
-    ) -> Self {
+    fn new(descriptors: &'a Descriptors, dynamic_transforms: &'a DynamicTransforms) -> Self {
         let transforms = Self::new_transforms(descriptors, dynamic_transforms);
 
         // DirectX does support drawing lines, but it's very inconsistent.
@@ -513,15 +449,7 @@ impl<'a> WgpuCommandHandler<'a> {
 
         Self {
             descriptors,
-            quality,
-            width,
-            height,
-            nearest_layer,
-            meshes,
-            staging_belt,
             dynamic_transforms,
-            draw_encoder,
-            texture_pool,
             emulate_lines,
 
             result: vec![],
@@ -610,104 +538,21 @@ impl<'a> WgpuCommandHandler<'a> {
 
 impl CommandHandler for WgpuCommandHandler<'_> {
     fn blend(&mut self, commands: CommandList, blend_mode: RenderBlendMode) {
-        let mut surface = Surface::new(
-            self.descriptors,
-            self.quality,
-            self.width,
-            self.height,
-            wgpu::TextureFormat::Rgba8Unorm,
-        );
-        let target_layer = if let RenderBlendMode::Builtin(BlendMode::Layer) = &blend_mode {
-            LayerRef::Current
-        } else {
-            self.nearest_layer
-        };
-        let blend_type = BlendType::from(blend_mode);
-        let clear_color = blend_type.default_color();
-        let target = surface.draw_commands(
-            RenderTargetMode::FreshWithColor(clear_color),
-            self.descriptors,
-            self.meshes,
-            commands,
-            self.staging_belt,
-            self.dynamic_transforms,
-            self.draw_encoder,
-            target_layer,
-            self.texture_pool,
-        );
-        target.ensure_cleared(self.draw_encoder);
-
-        match blend_type {
-            BlendType::Trivial(blend_mode) => {
-                let transform = Transform {
-                    matrix: Matrix::scale(target.width() as f32, target.height() as f32),
-                    color_transform: Default::default(),
-                };
-                let texture = target.take_color_texture();
-                let bind_group =
-                    self.descriptors
-                        .device
-                        .create_bind_group(&wgpu::BindGroupDescriptor {
-                            layout: &self.descriptors.bind_layouts.bitmap,
-                            entries: &[
-                                wgpu::BindGroupEntry {
-                                    binding: 0,
-                                    resource: self
-                                        .descriptors
-                                        .quad
-                                        .texture_transforms
-                                        .as_entire_binding(),
-                                },
-                                wgpu::BindGroupEntry {
-                                    binding: 1,
-                                    resource: wgpu::BindingResource::TextureView(texture.view()),
-                                },
-                                wgpu::BindGroupEntry {
-                                    binding: 2,
-                                    resource: wgpu::BindingResource::Sampler(
-                                        self.descriptors.bitmap_samplers.get_sampler(false, false),
-                                    ),
-                                },
-                            ],
-                            label: None,
-                        });
-                self.add_to_current(
-                    transform.matrix,
-                    transform.color_transform,
-                    |transform_buffer| DrawCommand::RenderTexture {
-                        _texture: texture,
-                        binds: bind_group,
-                        transform_buffer,
-                        blend_mode,
-                    },
-                );
-            }
-            blend_type => {
-                if !self.current.is_empty() {
-                    self.result.push(Chunk::Draw(
-                        mem::take(&mut self.current),
-                        self.needs_stencil,
-                        mem::replace(
-                            &mut self.transforms,
-                            BufferBuilder::new_for_uniform(&self.descriptors.limits),
-                        ),
-                    ));
-                }
-                self.transforms
-                    .set_buffer_limit(self.dynamic_transforms.buffer.size());
-                let chunk_blend_mode = match blend_type {
-                    BlendType::Complex(complex) => ChunkBlendMode::Complex(complex),
-                    BlendType::Shader(shader) => ChunkBlendMode::Shader(shader),
-                    _ => unreachable!(),
-                };
-                self.result.push(Chunk::Blend(
-                    target.take_color_texture(),
-                    chunk_blend_mode,
-                    self.num_masks > 0,
-                ));
-                self.needs_stencil = self.num_masks > 0;
-            }
+        if !self.current.is_empty() {
+            self.result.push(Chunk::Draw(
+                mem::take(&mut self.current),
+                self.needs_stencil,
+                mem::replace(
+                    &mut self.transforms,
+                    BufferBuilder::new_for_uniform(&self.descriptors.limits),
+                ),
+            ));
+            self.transforms
+                .set_buffer_limit(self.dynamic_transforms.buffer.size());
         }
+        self.result
+            .push(Chunk::Blend(commands, self.num_masks > 0, blend_mode));
+        self.needs_stencil = self.num_masks > 0;
     }
 
     fn render_bitmap(


Test patch: