.version 6.5
.target sm_30
.address_size 64

.extern .func __assertfail
(
	.param .b64 __assertfail_param_0,
	.param .b64 __assertfail_param_1,
	.param .b32 __assertfail_param_2,
	.param .b64 __assertfail_param_3,
	.param .b64 __assertfail_param_4
);

.extern .func __assertfail
(
	.param .b64 __assertfail_param_0,
	.param .b64 __assertfail_param_1,
	.param .b32 __assertfail_param_2,
	.param .b64 __assertfail_param_3,
	.param .b64 __assertfail_param_4
);

.visible .entry assertfail(
	.param .u64 input,
	.param .u64 output
)
{
	.reg .u64 	    in_addr;
    .reg .u64 	    out_addr;
    .reg .u64 	    temp;
    .reg .u64 	    temp2;

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



	{
	.reg .b32 temp_param_reg;
	mov.u32   temp_param_reg, 0;
	// <end>}
	.param .b64 param0;
	st.param.b64	[param0+0], in_addr;
	.param .b64 param1;
	st.param.b64	[param1+0], in_addr;
	.param .b32 param2;
	st.param.b32	[param2+0], temp_param_reg;
	.param .b64 param3;
	st.param.b64	[param3+0], in_addr;
	.param .b64 param4;
	st.param.b64	[param4+0], in_addr;
	call.uni 
	__assertfail, 
	(
	param0, 
	param1, 
	param2, 
	param3, 
	param4
	);
	
	//{
	}

    ld.u64          temp, [in_addr];
	add.u64		    temp2, temp, 1;
    st.u64          [out_addr], temp2;
	ret;
}


.extern .func __assertfail
(
	.param .b64 __assertfail_param_0,
	.param .b64 __assertfail_param_1,
	.param .b32 __assertfail_param_2,
	.param .b64 __assertfail_param_3,
	.param .b64 __assertfail_param_4
);