slang-test/WGPU: unfilterable-float type sampler used to filter

[aleino-nv]

This issue was split from https://github.com/shader-slang/slang/issues/5173.

Affected tests under tests/compute (currently excluded):

1. func-param-legalize

Output from slang-test, for func-param-legalize:

...
WGPU error: Texture binding (group:0, binding:0) is TextureSampleType::UnfilterableFloat but used statically with a sampler (group:0, binding:1) that's SamplerBindingType::Filtering
 - While validating compute stage ([ShaderModule (unlabeled)], entryPoint: computeMain).
 - While calling [Device].CreateComputePipeline([ComputePipelineDescriptor]).

WGPU error: [Invalid ComputePipeline (unlabeled)] is invalid.
 - While encoding [ComputePassEncoder (unlabeled)].SetPipeline([Invalid ComputePipeline (unlabeled)]).
 - While finishing [CommandEncoder (unlabeled)].

WGPU error: [Invalid CommandBuffer] is invalid.
 - While calling [Queue].Submit([[Invalid CommandBuffer]])
...

Generated WGSL for slangc -o test.wgsl -target wgsl -stage compute -entry computeMain %SLANG_SRC%\tests\compute\func-param-legalize.slang:

@binding(0) @group(0) var diffuseMap_0 : texture_2d<f32>;

@binding(1) @group(0) var samplerState_0 : sampler;

@binding(2) @group(0) var<storage, read_write> outputBuffer_0 : array<f32>;

struct Param_0
{
     base_0 : f32,
};

fn run_0( p_0 : Param_0,  p_tex_0 : texture_2d<f32>,  p_samplerState_0 : sampler) -> vec4<f32>
{
    return (textureSampleLevel((p_tex_0), (p_samplerState_0), (vec2<f32>(0.0f)), (0.0f))) + p_0.base_0;
}

@compute
@workgroup_size(1, 1, 1)
fn computeMain(@builtin(global_invocation_id) dispatchThreadID_0 : vec3<u32>)
{
    var p_1 : Param_0;
    p_1.base_0 = -0.5f;
    var _S1 : vec4<f32> = run_0(p_1, diffuseMap_0, samplerState_0);
    outputBuffer_0[i32(0)] = _S1.x;
    outputBuffer_0[i32(1)] = _S1.y;
    outputBuffer_0[i32(2)] = _S1.z;
    outputBuffer_0[i32(3)] = _S1.w;
    return;
}

[aleino-nv]

Here is a description of unfilterable float sample type: https://www.w3.org/TR/webgpu/#dom-gputexturesampletype-unfilterable-float

@skallweitNV I guess we should look at what slang-RHI supplies. Seems to me like it should supply the float sample type in this case. Maybe some reflection data is wrong?

[aleino-nv]

Yeah it seems we just return unfilterable-float unconditionally.

https://github.com/shader-slang/slang-rhi/blob/84f8ce13f92e9538b2e15f03d6bdac5ad9af215a/src/wgpu/wgpu-shader-object-layout.cpp#L54

[jkwak-work]

As far as I know, textures are not available on compute stage with WGSL. It works only for the fragment shader.

[skallweitNV]

I would be super surprised if textures are not available in compute stage ...

[aleino-nv]

...is TextureSampleType::UnfilterableFloat but used statically with a sampler (group:0, binding:1) that's SamplerBindingType::Filtering

Oh, I think possibly the fix here is to make the sampler not be of type filtering -- I don't know where it would get derivatives and such generally used for filtering, in a compute shader.

[jkwak-work]

I was surprised too but may be I am reading it wrong. Please see it for your self, here

16.7.8. textureSample Samples a texture.

Must only be used in a fragment shader stage.

[aleino-nv]

I was surprised too but may be I am reading it wrong. Please see it for your self, here

16.7.8. textureSample Samples a texture. Must only be used in a fragment shader stage.

It could also be because you need derivative of screen-to-triangle transformation in order to pick MIP level.

https://www.w3.org/TR/WGSL/#texturesamplelevel seems to be fine. (Check compute.toys samples)

[jkwak-work]

Oh. Yes, that must be it. It makes sense now