| 283 | |
| 284 | |
| 285 | void Run( |
| 286 | KernelContext &context, |
| 287 | const OutListGPU<OutputType, ndims> &out, |
| 288 | const span<InListGPU<InputType, ndims>> &ins) { |
| 289 | paste::FillPointers(make_span(sample_descriptors_), make_span(patch_descriptors_), out, ins); |
| 290 | |
| 291 | SampleDesc *samples_gpu; |
| 292 | PatchDesc *patches_gpu; |
| 293 | BlockDesc *blocks_gpu; |
| 294 | |
| 295 | std::tie(samples_gpu, patches_gpu, blocks_gpu) = context.scratchpad->ToContiguousGPU( |
| 296 | context.gpu.stream, sample_descriptors_, patch_descriptors_, block_setup_.Blocks()); |
| 297 | |
| 298 | dim3 patch_dim = block_setup_.GridDim(); |
| 299 | dim3 block_dim = block_setup_.BlockDim(); |
| 300 | auto stream = context.gpu.stream; |
| 301 | |
| 302 | paste::PasteKernel<<<patch_dim, block_dim, 0, stream>>>( |
| 303 | samples_gpu, patches_gpu, blocks_gpu); |
| 304 | CUDA_CALL(cudaGetLastError()); |
| 305 | } |
| 306 | }; |
| 307 | |
| 308 | } // namespace kernels |
nothing calls this directly
no test coverage detected