Skip to content

Instantly share code, notes, and snippets.

@karolherbst

karolherbst/llvm-16 Secret

Created February 14, 2024 22:37
Show Gist options
  • Save karolherbst/67bc4a4e2bd62714b92899c2ae4e8cd7 to your computer and use it in GitHub Desktop.
Save karolherbst/67bc4a4e2bd62714b92899c2ae4e8cd7 to your computer and use it in GitHub Desktop.
Download ZIP
Raw
llvm-16
shader: MESA_SHADER_KERNEL
source_sha1: {0x8bd887f1, 0xa3a04835, 0xaf69089c, 0x96218356, 0x2bf875e4}
internal: false
workgroup_size: 0, 0, 0 (variable)
stage: 14
next_stage: 0
float_controls_execution_mode: 0x00038011
subgroup_size: 0
writes_memory: true
ptr_size: 64
inputs: 0
outputs: 0
uniforms: 16
scratch: 20
constants: 20
decl_var uniform INTERP_MODE_NONE none uint64_t #0 (0, 0, 0)
decl_var uniform INTERP_MODE_NONE none uint64_t constant_buffer_addr (2, 8, 0)
decl_var constant INTERP_MODE_NONE none struct.GFX125_3DSTATE_VERTEX_BUFFERS __const.write_3DSTATE_VERTEX_BUFFERS.v = { { 0x00000003 }, { 0x00000008 }, { 0x00000000 }, { 0x00000003 }, { 0x00000003 } }
decl_function __wrapped_write_3DSTATE_VERTEX_BUFFERS (0 params)
impl __wrapped_write_3DSTATE_VERTEX_BUFFERS {
decl_var INTERP_MODE_NONE none struct.GFX125_3DSTATE_VERTEX_BUFFERS v
block b0: // preds:
64 %0 = deref_var &#0 (uniform uint64_t)
64 %1 = @load_deref (%0) (access=none)
64 %2 = deref_var &v (function_temp struct.GFX125_3DSTATE_VERTEX_BUFFERS)
64 %3 = deref_var &__const.write_3DSTATE_VERTEX_BUFFERS.v (constant struct.GFX125_3DSTATE_VERTEX_BUFFERS)
64 %11 = deref_struct &%3->field0 (constant uint) // &__const.write_3DSTATE_VERTEX_BUFFERS.v.field0
64 %12 = deref_struct &%2->field0 (function_temp uint) // &v.field0
32 %21 = @load_deref (%11) (access=none)
@store_deref (%12, %21) (wrmask=x, access=none)
64 %13 = deref_struct &%3->field1 (constant uint) // &__const.write_3DSTATE_VERTEX_BUFFERS.v.field1
64 %14 = deref_struct &%2->field1 (function_temp uint) // &v.field1
32 %22 = @load_deref (%13) (access=none)
@store_deref (%14, %22) (wrmask=x, access=none)
64 %15 = deref_struct &%3->field2 (constant uint) // &__const.write_3DSTATE_VERTEX_BUFFERS.v.field2
64 %16 = deref_struct &%2->field2 (function_temp uint) // &v.field2
32 %23 = @load_deref (%15) (access=none)
@store_deref (%16, %23) (wrmask=x, access=none)
64 %17 = deref_struct &%3->field3 (constant uint) // &__const.write_3DSTATE_VERTEX_BUFFERS.v.field3
64 %18 = deref_struct &%2->field3 (function_temp uint) // &v.field3
32 %24 = @load_deref (%17) (access=none)
@store_deref (%18, %24) (wrmask=x, access=none)
64 %19 = deref_struct &%3->field4 (constant uint) // &__const.write_3DSTATE_VERTEX_BUFFERS.v.field4
64 %20 = deref_struct &%2->field4 (function_temp uint) // &v.field4
32 %25 = @load_deref (%19) (access=none)
@store_deref (%20, %25) (wrmask=x, access=none)
64 %7 = deref_struct &%2->field0 (function_temp uint) // &v.field0
32 %9 = @load_deref (%7) (access=none)
64 %10 = deref_cast (uint *)%1 (global uint) (ptr_stride=4, align_mul=4, align_offset=0)
@store_deref (%10, %9) (wrmask=x, access=none)
// succs: b1
block b1:
}
nir_lower_vars_to_ssa
shader: MESA_SHADER_KERNEL
source_sha1: {0x8bd887f1, 0xa3a04835, 0xaf69089c, 0x96218356, 0x2bf875e4}
internal: false
workgroup_size: 0, 0, 0 (variable)
stage: 14
next_stage: 0
float_controls_execution_mode: 0x00038011
subgroup_size: 0
writes_memory: true
ptr_size: 64
inputs: 0
outputs: 0
uniforms: 16
scratch: 20
constants: 20
decl_var uniform INTERP_MODE_NONE none uint64_t #0 (0, 0, 0)
decl_var uniform INTERP_MODE_NONE none uint64_t constant_buffer_addr (2, 8, 0)
decl_var constant INTERP_MODE_NONE none struct.GFX125_3DSTATE_VERTEX_BUFFERS __const.write_3DSTATE_VERTEX_BUFFERS.v = { { 0x00000003 }, { 0x00000008 }, { 0x00000000 }, { 0x00000003 }, { 0x00000003 } }
decl_function __wrapped_write_3DSTATE_VERTEX_BUFFERS (0 params)
impl __wrapped_write_3DSTATE_VERTEX_BUFFERS {
decl_var INTERP_MODE_NONE none struct.GFX125_3DSTATE_VERTEX_BUFFERS v
block b0: // preds:
64 %0 = deref_var &#0 (uniform uint64_t)
64 %1 = @load_deref (%0) (access=none)
64 %2 = deref_var &v (function_temp struct.GFX125_3DSTATE_VERTEX_BUFFERS)
64 %3 = deref_var &__const.write_3DSTATE_VERTEX_BUFFERS.v (constant struct.GFX125_3DSTATE_VERTEX_BUFFERS)
64 %11 = deref_struct &%3->field0 (constant uint) // &__const.write_3DSTATE_VERTEX_BUFFERS.v.field0
64 %12 = deref_struct &%2->field0 (function_temp uint) // &v.field0
32 %21 = @load_deref (%11) (access=none)
64 %13 = deref_struct &%3->field1 (constant uint) // &__const.write_3DSTATE_VERTEX_BUFFERS.v.field1
64 %14 = deref_struct &%2->field1 (function_temp uint) // &v.field1
32 %22 = @load_deref (%13) (access=none)
64 %15 = deref_struct &%3->field2 (constant uint) // &__const.write_3DSTATE_VERTEX_BUFFERS.v.field2
64 %16 = deref_struct &%2->field2 (function_temp uint) // &v.field2
32 %23 = @load_deref (%15) (access=none)
64 %17 = deref_struct &%3->field3 (constant uint) // &__const.write_3DSTATE_VERTEX_BUFFERS.v.field3
64 %18 = deref_struct &%2->field3 (function_temp uint) // &v.field3
32 %24 = @load_deref (%17) (access=none)
64 %19 = deref_struct &%3->field4 (constant uint) // &__const.write_3DSTATE_VERTEX_BUFFERS.v.field4
64 %20 = deref_struct &%2->field4 (function_temp uint) // &v.field4
32 %25 = @load_deref (%19) (access=none)
64 %7 = deref_struct &%2->field0 (function_temp uint) // &v.field0
32 %26 = mov %21
64 %10 = deref_cast (uint *)%1 (global uint) (ptr_stride=4, align_mul=4, align_offset=0)
@store_deref (%10, %26) (wrmask=x, access=none)
// succs: b1
block b1:
}
nir_lower_alu
nir_opt_phi_precision
nir_opt_algebraic
nir_opt_if
nir_opt_dead_cf
nir_opt_remove_phis
nir_opt_peephole_select
nir_lower_vec3_to_vec4
nir_opt_loop_unroll
nir_copy_prop
shader: MESA_SHADER_KERNEL
source_sha1: {0x8bd887f1, 0xa3a04835, 0xaf69089c, 0x96218356, 0x2bf875e4}
internal: false
workgroup_size: 0, 0, 0 (variable)
stage: 14
next_stage: 0
float_controls_execution_mode: 0x00038011
subgroup_size: 0
writes_memory: true
ptr_size: 64
inputs: 0
outputs: 0
uniforms: 16
scratch: 20
constants: 20
decl_var uniform INTERP_MODE_NONE none uint64_t #0 (0, 0, 0)
decl_var uniform INTERP_MODE_NONE none uint64_t constant_buffer_addr (2, 8, 0)
decl_var constant INTERP_MODE_NONE none struct.GFX125_3DSTATE_VERTEX_BUFFERS __const.write_3DSTATE_VERTEX_BUFFERS.v = { { 0x00000003 }, { 0x00000008 }, { 0x00000000 }, { 0x00000003 }, { 0x00000003 } }
decl_function __wrapped_write_3DSTATE_VERTEX_BUFFERS (0 params)
impl __wrapped_write_3DSTATE_VERTEX_BUFFERS {
decl_var INTERP_MODE_NONE none struct.GFX125_3DSTATE_VERTEX_BUFFERS v
block b0: // preds:
64 %0 = deref_var &#0 (uniform uint64_t)
64 %1 = @load_deref (%0) (access=none)
64 %2 = deref_var &v (function_temp struct.GFX125_3DSTATE_VERTEX_BUFFERS)
64 %3 = deref_var &__const.write_3DSTATE_VERTEX_BUFFERS.v (constant struct.GFX125_3DSTATE_VERTEX_BUFFERS)
64 %4 = deref_struct &%3->field0 (constant uint) // &__const.write_3DSTATE_VERTEX_BUFFERS.v.field0
64 %5 = deref_struct &%2->field0 (function_temp uint) // &v.field0
32 %6 = @load_deref (%4) (access=none)
64 %7 = deref_struct &%3->field1 (constant uint) // &__const.write_3DSTATE_VERTEX_BUFFERS.v.field1
64 %8 = deref_struct &%2->field1 (function_temp uint) // &v.field1
32 %9 = @load_deref (%7) (access=none)
64 %10 = deref_struct &%3->field2 (constant uint) // &__const.write_3DSTATE_VERTEX_BUFFERS.v.field2
64 %11 = deref_struct &%2->field2 (function_temp uint) // &v.field2
32 %12 = @load_deref (%10) (access=none)
64 %13 = deref_struct &%3->field3 (constant uint) // &__const.write_3DSTATE_VERTEX_BUFFERS.v.field3
64 %14 = deref_struct &%2->field3 (function_temp uint) // &v.field3
32 %15 = @load_deref (%13) (access=none)
64 %16 = deref_struct &%3->field4 (constant uint) // &__const.write_3DSTATE_VERTEX_BUFFERS.v.field4
64 %17 = deref_struct &%2->field4 (function_temp uint) // &v.field4
32 %18 = @load_deref (%16) (access=none)
64 %19 = deref_struct &%2->field0 (function_temp uint) // &v.field0
64 %21 = deref_cast (uint *)%1 (global uint) (ptr_stride=4, align_mul=4, align_offset=0)
@store_deref (%21, %6) (wrmask=x, access=none)
// succs: b1
block b1:
}
nir_opt_copy_prop_vars
nir_opt_dead_write_vars
nir_lower_alu_to_scalar
nir_lower_phis_to_scalar
nir_opt_deref
nir_opt_memcpy
nir_opt_dce
shader: MESA_SHADER_KERNEL
source_sha1: {0x8bd887f1, 0xa3a04835, 0xaf69089c, 0x96218356, 0x2bf875e4}
internal: false
workgroup_size: 0, 0, 0 (variable)
stage: 14
next_stage: 0
float_controls_execution_mode: 0x00038011
subgroup_size: 0
writes_memory: true
ptr_size: 64
inputs: 0
outputs: 0
uniforms: 16
scratch: 20
constants: 20
decl_var uniform INTERP_MODE_NONE none uint64_t #0 (0, 0, 0)
decl_var uniform INTERP_MODE_NONE none uint64_t constant_buffer_addr (2, 8, 0)
decl_var constant INTERP_MODE_NONE none struct.GFX125_3DSTATE_VERTEX_BUFFERS __const.write_3DSTATE_VERTEX_BUFFERS.v = { { 0x00000003 }, { 0x00000008 }, { 0x00000000 }, { 0x00000003 }, { 0x00000003 } }
decl_function __wrapped_write_3DSTATE_VERTEX_BUFFERS (0 params)
impl __wrapped_write_3DSTATE_VERTEX_BUFFERS {
decl_var INTERP_MODE_NONE none struct.GFX125_3DSTATE_VERTEX_BUFFERS v
block b0: // preds:
64 %0 = deref_var &#0 (uniform uint64_t)
64 %1 = @load_deref (%0) (access=none)
64 %3 = deref_var &__const.write_3DSTATE_VERTEX_BUFFERS.v (constant struct.GFX125_3DSTATE_VERTEX_BUFFERS)
64 %4 = deref_struct &%3->field0 (constant uint) // &__const.write_3DSTATE_VERTEX_BUFFERS.v.field0
32 %6 = @load_deref (%4) (access=none)
64 %21 = deref_cast (uint *)%1 (global uint) (ptr_stride=4, align_mul=4, align_offset=0)
@store_deref (%21, %6) (wrmask=x, access=none)
// succs: b1
block b1:
}
Raw
llvm-17
shader: MESA_SHADER_KERNEL
source_sha1: {0x945f51b1, 0xcf3b29d0, 0x7186ab9c, 0xa3d0bcb8, 0x0e446417}
internal: false
workgroup_size: 0, 0, 0 (variable)
stage: 14
next_stage: 0
float_controls_execution_mode: 0x00038011
subgroup_size: 0
writes_memory: true
ptr_size: 64
inputs: 0
outputs: 0
uniforms: 16
scratch: 20
constants: 20
decl_var uniform INTERP_MODE_NONE none uint64_t #0 (0, 0, 0)
decl_var uniform INTERP_MODE_NONE none uint64_t constant_buffer_addr (2, 8, 0)
decl_var constant INTERP_MODE_NONE none struct.GFX125_3DSTATE_VERTEX_BUFFERS __const.write_3DSTATE_VERTEX_BUFFERS.v = { { 0x00000003 }, { 0x00000008 }, { 0x00000000 }, { 0x00000003 }, { 0x00000003 } }
decl_function __wrapped_write_3DSTATE_VERTEX_BUFFERS (0 params)
impl __wrapped_write_3DSTATE_VERTEX_BUFFERS {
decl_var INTERP_MODE_NONE none struct.GFX125_3DSTATE_VERTEX_BUFFERS v
block b0: // preds:
64 %0 = deref_var &#0 (uniform uint64_t)
64 %1 = @load_deref (%0) (access=none)
64 %3 = deref_var &__const.write_3DSTATE_VERTEX_BUFFERS.v (constant struct.GFX125_3DSTATE_VERTEX_BUFFERS)
64 %4 = deref_var &v (function_temp struct.GFX125_3DSTATE_VERTEX_BUFFERS)
64 %8 = deref_struct &%3->field0 (constant uint) // &__const.write_3DSTATE_VERTEX_BUFFERS.v.field0
64 %9 = deref_struct &%4->field0 (function_temp uint) // &v.field0
32 %18 = @load_deref (%8) (access=none)
@store_deref (%9, %18) (wrmask=x, access=none)
64 %10 = deref_struct &%3->field1 (constant uint) // &__const.write_3DSTATE_VERTEX_BUFFERS.v.field1
64 %11 = deref_struct &%4->field1 (function_temp uint) // &v.field1
32 %19 = @load_deref (%10) (access=none)
@store_deref (%11, %19) (wrmask=x, access=none)
64 %12 = deref_struct &%3->field2 (constant uint) // &__const.write_3DSTATE_VERTEX_BUFFERS.v.field2
64 %13 = deref_struct &%4->field2 (function_temp uint) // &v.field2
32 %20 = @load_deref (%12) (access=none)
@store_deref (%13, %20) (wrmask=x, access=none)
64 %14 = deref_struct &%3->field3 (constant uint) // &__const.write_3DSTATE_VERTEX_BUFFERS.v.field3
64 %15 = deref_struct &%4->field3 (function_temp uint) // &v.field3
32 %21 = @load_deref (%14) (access=none)
@store_deref (%15, %21) (wrmask=x, access=none)
64 %16 = deref_struct &%3->field4 (constant uint) // &__const.write_3DSTATE_VERTEX_BUFFERS.v.field4
64 %17 = deref_struct &%4->field4 (function_temp uint) // &v.field4
32 %22 = @load_deref (%16) (access=none)
@store_deref (%17, %22) (wrmask=x, access=none)
64 %5 = deref_cast (uint *)%4 (function_temp uint) (ptr_stride=4, align_mul=0, align_offset=0)
32 %6 = @load_deref (%5) (access=none)
64 %7 = deref_cast (uint *)%1 (global uint) (ptr_stride=4, align_mul=4, align_offset=0)
@store_deref (%7, %6) (wrmask=x, access=none)
// succs: b1
block b1:
}
nir_lower_vars_to_ssa
nir_lower_alu
nir_opt_phi_precision
nir_opt_algebraic
nir_opt_if
nir_opt_dead_cf
nir_opt_remove_phis
nir_opt_peephole_select
nir_lower_vec3_to_vec4
nir_opt_loop_unroll
nir_copy_prop
nir_opt_copy_prop_vars
nir_opt_dead_write_vars
nir_lower_alu_to_scalar
nir_lower_phis_to_scalar
nir_opt_deref
nir_opt_memcpy
nir_opt_dce
nir_opt_undef
nir_opt_constant_folding
shader: MESA_SHADER_KERNEL
source_sha1: {0x945f51b1, 0xcf3b29d0, 0x7186ab9c, 0xa3d0bcb8, 0x0e446417}
internal: false
workgroup_size: 0, 0, 0 (variable)
stage: 14
next_stage: 0
float_controls_execution_mode: 0x00038011
subgroup_size: 0
writes_memory: true
ptr_size: 64
inputs: 0
outputs: 0
uniforms: 16
scratch: 20
constants: 20
decl_var uniform INTERP_MODE_NONE none uint64_t #0 (0, 0, 0)
decl_var uniform INTERP_MODE_NONE none uint64_t constant_buffer_addr (2, 8, 0)
decl_var constant INTERP_MODE_NONE none struct.GFX125_3DSTATE_VERTEX_BUFFERS __const.write_3DSTATE_VERTEX_BUFFERS.v = { { 0x00000003 }, { 0x00000008 }, { 0x00000000 }, { 0x00000003 }, { 0x00000003 } }
decl_function __wrapped_write_3DSTATE_VERTEX_BUFFERS (0 params)
impl __wrapped_write_3DSTATE_VERTEX_BUFFERS {
decl_var INTERP_MODE_NONE none struct.GFX125_3DSTATE_VERTEX_BUFFERS v
block b0: // preds:
64 %0 = deref_var &#0 (uniform uint64_t)
64 %1 = @load_deref (%0) (access=none)
64 %2 = deref_var &__const.write_3DSTATE_VERTEX_BUFFERS.v (constant struct.GFX125_3DSTATE_VERTEX_BUFFERS)
64 %3 = deref_var &v (function_temp struct.GFX125_3DSTATE_VERTEX_BUFFERS)
64 %4 = deref_struct &%2->field0 (constant uint) // &__const.write_3DSTATE_VERTEX_BUFFERS.v.field0
64 %5 = deref_struct &%3->field0 (function_temp uint) // &v.field0
32 %22 = load_const (0x00000003)
@store_deref (%5, %22 (0x3)) (wrmask=x, access=none)
64 %7 = deref_struct &%2->field1 (constant uint) // &__const.write_3DSTATE_VERTEX_BUFFERS.v.field1
64 %8 = deref_struct &%3->field1 (function_temp uint) // &v.field1
32 %23 = load_const (0x00000008)
@store_deref (%8, %23 (0x8)) (wrmask=x, access=none)
64 %10 = deref_struct &%2->field2 (constant uint) // &__const.write_3DSTATE_VERTEX_BUFFERS.v.field2
64 %11 = deref_struct &%3->field2 (function_temp uint) // &v.field2
32 %24 = load_const (0x00000000)
@store_deref (%11, %24 (0x0)) (wrmask=x, access=none)
64 %13 = deref_struct &%2->field3 (constant uint) // &__const.write_3DSTATE_VERTEX_BUFFERS.v.field3
64 %14 = deref_struct &%3->field3 (function_temp uint) // &v.field3
32 %25 = load_const (0x00000003)
@store_deref (%14, %25 (0x3)) (wrmask=x, access=none)
64 %16 = deref_struct &%2->field4 (constant uint) // &__const.write_3DSTATE_VERTEX_BUFFERS.v.field4
64 %17 = deref_struct &%3->field4 (function_temp uint) // &v.field4
32 %26 = load_const (0x00000003)
@store_deref (%17, %26 (0x3)) (wrmask=x, access=none)
64 %19 = deref_cast (uint *)%3 (function_temp uint) (ptr_stride=4, align_mul=0, align_offset=0)
32 %20 = @load_deref (%19) (access=none)
64 %21 = deref_cast (uint *)%1 (global uint) (ptr_stride=4, align_mul=4, align_offset=0)
@store_deref (%21, %20) (wrmask=x, access=none)
// succs: b1
block b1:
}
Sign up for free to join this conversation on GitHub. Already have an account? Sign in to comment