Much as the kernel layer insulates the user from the programming details of a series of specific tasks, the module layer insulates the user from the implementation details of the module of -RTM, which contains forward extrapolation, wavefield reconstruction, backward extrapolation and imaging. Each of them is made up of several kernel functions and streams. A stream is defined by creating a stream object using cudaStreamCreate() and specifying it as the stream parameter to a sequence of kernel launches and host-device memory copies. Streams are released by calling cudaStreamDestroy(), which waits for all preceding commands in the given stream to complete before destroying the stream and returning control to the host thread. THe forward module cuda_visco_PSM_2d_forward() in cu-RTM is designed in a splitting fashion and called by the main function to conduct forward wavefield extrapolation. As wavefield reconstruction, backward extrapolation and imaging are conducted during time-reversal simulation, these three modules can be merged into one module cuda_visco_PSM_2d_backward(). Both forward and backward modules are presented by brief codes in Appendix A.