14:11 <alyssa> lina: follow up from wednesday https://gitlab.freedesktop.org/asahi/mesa/-/merge_requests/88

14:11 <alyssa> because I decided I'd had enough of CTS crunch time and wanted to write a silly optimization pass just for fun ;)

14:12 <alyssa> please test with Darwinia since clearly I'm no good at testing discard, lolz

14:51 <alyssa> --

14:51 <alyssa> It's nice, retesting Dolphin and seeing perf has creeped up since last time

14:51 <alyssa> not by a lot, but a little

16:32 <alyssa> and with a zero extension has flags 3, second source

16:45 <alyssa> for sample shading + fs side effects, Apple wraps device_store instructions in `if ((1 << sample_id) & sr60l != 0)`

16:46 <alyssa> for sample shading + fs side effects + discard..

16:47 <alyssa> + early frag

16:47 <alyssa> Each iteration of the sample loop is wrapped in `if ((1 << sample_id) & sr60l != 0)`

16:48 <alyssa> um this is probably invoking undefined behaviour isn't it lol

16:50 <alyssa> no this is ok

16:50 <alyssa> Inside that if, it does

16:50 <alyssa> sample_mask(1 << sample_id, ~0)

16:50 <alyssa> if (discard condition) {

16:51 <alyssa> sample_mask(1 << sample_id, 0)

16:51 <alyssa> }

16:51 <alyssa> store

16:51 <alyssa> sample_mask (1<<sample_id, ~0)

16:51 <alyssa> yeah no this seems wrong

16:55 <alyssa> in principle, this is saying "run depth/stencil tests, then discard" as required by the spec

16:55 <alyssa> but how does the store get masked?

16:56 <alyssa> isn't it running tets twice?

16:58 <alyssa> not sure what to make of this

17:01 <alyssa> maybe a metal bug, who knows

17:02 <alyssa> /System/Library/PrivateFrameworks/GPUCompiler.framework/Versions/31001/Libraries/lib/clang/31001.523/include/metal/metal_texture:7168:3: error: static_assert failed due to requirement '_is_any_of((metal::access)2, access::read)' "Multisample textures must have access qualifier access::read"

17:02 <alyssa> Oooooooof

17:03 <alyssa> Ughhhhh

17:03 <alyssa> So does the hardware support multisampled images? >:

17:03 <alyssa> cause if not i'm going to be annoyed

17:41 <alyssa> should be straightforward to lower with a similar technique to texture atomics, but ugh