Signed modulo treated as unsigned on wgpu vulkan running on nvidia rtx 5080.
Reproducer:
@group(0) @binding(0) var<storage, read_write> output: i32;

@compute @workgroup_size(1)
fn main(@Builtin(global_invocation_id) id: vec3) {
var a: i32 = i32(id.x) - 1i;
output = a % 2i;
}

Output is calculated as 1 on vulkan (nvidia) for thread 0. On AMD Radeon integrated graphics vulkan, dx12 nvidia/amd, and dawn dx12 and vulkan on both nvidia and amd it is correctly calculated as -1.

Looking at the spv output of naga, it seems it (correctly) uses OpSRem. Meanwhile tint uses OpSDiv to compute lhs - (lhs/rhs) * rhs. This may be allowing it to bypass a driver issue which could explain why this works in dawn.

Run on:
Windows 11
9800x3d
rtx 5080
wgpu 29.0.3
Nvidia driver 596.49