ptx/src/test/spirv_run/stateful_ld_st_simple.ptx


1
2
3
4
5
6
7
8
9
10
11
12
13
14
15
16
17
18
19
20
21
22
23
24
25

.version 6.5
.target sm_30
.address_size 64

.visible .entry stateful_ld_st_simple(
    .param .u64 input,
    .param .u64 output
)
{
    .reg .u64           in_addr;
    .reg .u64           out_addr;
    .reg .u64           in_addr2;
    .reg .u64           out_addr2;
    .reg .u64           temp;

    ld.param.u64        in_addr, [input];
    ld.param.u64        out_addr, [output];

    cvta.to.global.u64  in_addr2, in_addr;
    cvta.to.global.u64  out_addr2, out_addr;

    ld.global.u64       temp, [in_addr2];
    st.global.u64       [out_addr2], temp;
    ret;
}