MSAA + copy_texture_to_buffer

[dmtrKovalenko]

First of all thanks for the great project!
Can not find any information about how I can implement the copying texture to buffer after multi-sampling antialiasing. I am following this tutorial to copy the texture to the buffer and then use it for my own needs. But it omits chaotic results when adding

     multisample: wgpu::MultisampleState {
            count: 4,
            mask: !0,
            alpha_to_coverage_enabled: false,
     },

Here is a result of pretty simple color + text rendering with count: 1

And here is what I am getting with count: 4

But when I am using winit and their default texture and frame and it works perfectly.

[kvark]

You shouldn't be able to copy MSAA texture into a buffer, at all, if I recall correctly. The fact wgpu allows it now is a bug.
What you can do instead is to run a compute shader, which does textureLoad() from the MSAA texture and stores data into a buffer (bound as a read/write storage buffer).

[dmtrKovalenko]

I am sorry, but I actually didn't get what you said. Spend some time investigating how begin_compute_pass works and I am not sure it is something I need. My task is actually to render vertices and indexes with MSAA and then get the buffer from the GPU. Is it possible without begin_render_pass?

[kvark]

Let me answer this in greater detail. First (and most important) question is - do you really need to read MSAA data, as opposed to reading pixels of the MSAA resolve?

I need MSAA data

A multisampled (MSAA) texture contains multiple values per pixel. The exact binary representation of this form is not a part of the standard - it's up to the driver to decide how to store them. Generally, drivers would want to do anything other than storing (sample0,sample1,sample2,...) because writing and reading all those values every time a pixel is touched is too expensive.

As the layout of each pixel is not disclosed to us, there is no point in trying to copy that data directly into a buffer (i.e. with copy_texture_to_buffer). We are going to forbid this with validation, and the fact that it's missing is just a bug right now.

If you expect a specific representation of MSAA in the buffer, you can get it manually by running a compute shader that loads the texture data and puts it into a buffer. The shader could look something like this:

@group(0) @binding(0)
var src_texture: texture_multisampled_2d<f32>;
@group(0) @binding(1)
var<storage, read_write> dst_buffer: array<vec4<f32>>;

@stage(compute) @workgroup_size(64, 1, 1)
fn main(
  @builtin(global_invocation_id) global_id: vec3<u32>,
  @builtin(local_invocation_index) local_index: u32,
  @builtin(workgroup_id) group_id: vec3<u32>,
  @builtin(num_workgroups) group_count: vec3<u32>,
) {
  let dimensions = vec2<u32>(textureDimensions(src_texture));
  if (global_id.x < dimensions.x && global_id.y < dimensions.y) {
    let group_index = group_id.x  + group_id.y * group_count.x + group_id.z * group_count.x * group_count.y;
    let global_index = group_index * 64 + local_index;
    // consider `global_id.z` to be the sample index
    dst_buffer[global_index] = textureLoad(src_texture, vec2<i32>(global_id.xy), i32(global_id.z));
  }
}

You'd need to set up the pipeline layout for this with the texture binding and a writable storage buffer. And you'd need to invoke it with the appropriate group counts. Something like this:

pass.set_bind_group(0, bind_group_with_texture_and_buffer);
pass.dispatch(texture_width/64, texture_height, 1);

Warning: this code was written by hand, and was not tested.

I just need smooth edges

This is what people generally need from MSAA. In this case, you don't actually need the MSAA samples being read. You need to have the HW to resolve those samples and produce a non-MSAA texture. Check the msaa-line example for the reference.

Basically, you'd create both MSAA and non-MSAA versions of the texture. Then in the render pass you'd point the resolve_target to the non-MSAA version, and set the store: false to avoid writing to the MSAA (as an optimization). At the end of the pass, the HW will resolve the texture and write it down for you.

[dmtrKovalenko]

Thanks for such a detailed explanation!