Home
last modified time | relevance | path

Searched refs:reduceBlock (Results 1 – 24 of 24) sorted by relevance

/dports/science/liggghts/LIGGGHTS-PUBLIC-3.8.0-26-g6e873439/lib/cuda/
H A Dcuda_pair_virial_kernel_nc.cu32 reduceBlock(shared); in PairVirialCompute_A_Kernel()
36 reduceBlock(shared); in PairVirialCompute_A_Kernel()
42 reduceBlock(shared + 0 * blockDim.x); in PairVirialCompute_A_Kernel()
43 reduceBlock(shared + 1 * blockDim.x); in PairVirialCompute_A_Kernel()
44 reduceBlock(shared + 2 * blockDim.x); in PairVirialCompute_A_Kernel()
45 reduceBlock(shared + 3 * blockDim.x); in PairVirialCompute_A_Kernel()
46 reduceBlock(shared + 4 * blockDim.x); in PairVirialCompute_A_Kernel()
47 reduceBlock(shared + 5 * blockDim.x); in PairVirialCompute_A_Kernel()
90 reduceBlock(sharedmem); in MY_AP()
H A Dcompute_temp_cuda_kernel.cu42 reduceBlock(sharedmem); in Cuda_ComputeTempCuda_Scalar_Kernel()
75 reduceBlock(sharedmem); in Cuda_ComputeTempCuda_Vector_Kernel()
76 reduceBlock(&sharedmem[blockDim.x]); in Cuda_ComputeTempCuda_Vector_Kernel()
77 reduceBlock(&sharedmem[2 * blockDim.x]); in Cuda_ComputeTempCuda_Vector_Kernel()
78 reduceBlock(&sharedmem[3 * blockDim.x]); in Cuda_ComputeTempCuda_Vector_Kernel()
79 reduceBlock(&sharedmem[4 * blockDim.x]); in Cuda_ComputeTempCuda_Vector_Kernel()
80 reduceBlock(&sharedmem[5 * blockDim.x]); in Cuda_ComputeTempCuda_Vector_Kernel()
109 reduceBlock(sharedmem); in Cuda_ComputeTempCuda_Reduce_Kernel()
H A Dcompute_temp_partial_cuda_kernel.cu42 reduceBlock(sharedmem); in Cuda_ComputeTempPartialCuda_Scalar_Kernel()
75 reduceBlock(sharedmem); in Cuda_ComputeTempPartialCuda_Vector_Kernel()
76 reduceBlock(&sharedmem[blockDim.x]); in Cuda_ComputeTempPartialCuda_Vector_Kernel()
77 reduceBlock(&sharedmem[2 * blockDim.x]); in Cuda_ComputeTempPartialCuda_Vector_Kernel()
78 reduceBlock(&sharedmem[3 * blockDim.x]); in Cuda_ComputeTempPartialCuda_Vector_Kernel()
79 reduceBlock(&sharedmem[4 * blockDim.x]); in Cuda_ComputeTempPartialCuda_Vector_Kernel()
80 reduceBlock(&sharedmem[5 * blockDim.x]); in Cuda_ComputeTempPartialCuda_Vector_Kernel()
109 reduceBlock(sharedmem); in Cuda_ComputeTempPartialCuda_Reduce_Kernel()
H A Dfix_addforce_cuda_kernel.cu50 reduceBlock(sharedmem); in Cuda_FixAddForceCuda_PostForce_Kernel()
51 reduceBlock(&sharedmem[blockDim.x]); in Cuda_FixAddForceCuda_PostForce_Kernel()
52 reduceBlock(&sharedmem[2 * blockDim.x]); in Cuda_FixAddForceCuda_PostForce_Kernel()
53 reduceBlock(&sharedmem[3 * blockDim.x]); in Cuda_FixAddForceCuda_PostForce_Kernel()
81 reduceBlock(sharedmem); in reduce_foriginal()
H A Dfix_aveforce_cuda_kernel.cu43 reduceBlock(sharedmem); in Cuda_FixAveForceCuda_PostForce_FOrg_Kernel()
44 reduceBlock(&sharedmem[blockDim.x]); in Cuda_FixAveForceCuda_PostForce_FOrg_Kernel()
45 reduceBlock(&sharedmem[2 * blockDim.x]); in Cuda_FixAveForceCuda_PostForce_FOrg_Kernel()
46 reduceBlock(&sharedmem[3 * blockDim.x]); in Cuda_FixAveForceCuda_PostForce_FOrg_Kernel()
73 reduceBlock(sharedmem); in Cuda_FixAveForceCuda_reduce_foriginal()
H A Dfix_freeze_cuda_kernel.cu49 reduceBlock(sharedmem); in Cuda_FixFreezeCuda_PostForce_Kernel()
50 reduceBlock(&sharedmem[blockDim.x]); in Cuda_FixFreezeCuda_PostForce_Kernel()
51 reduceBlock(&sharedmem[2 * blockDim.x]); in Cuda_FixFreezeCuda_PostForce_Kernel()
77 reduceBlock(sharedmem); in Cuda_FixFreezeCuda_Reduce_FOriginal()
H A Dfix_set_force_cuda_kernel.cu48 reduceBlock(sharedmem); in Cuda_FixSetForceCuda_PostForce_Kernel()
49 reduceBlock(&sharedmem[blockDim.x]); in Cuda_FixSetForceCuda_PostForce_Kernel()
50 reduceBlock(&sharedmem[2 * blockDim.x]); in Cuda_FixSetForceCuda_PostForce_Kernel()
76 reduceBlock(sharedmem); in Cuda_FixSetForceCuda_Reduce_FOriginal()
H A Dpppm_cuda_kernel.cu30 __device__ void reduceBlock(float* data) in reduceBlock() function
49 __device__ void reduceBlock(double* data) in reduceBlock() function
259 reduceBlock(s_energy); in poisson_energy_kernel()
270 reduceBlock(s_energy); in poisson_energy_kernel()
285 reduceBlock(s_energy); in sum_energy_kernel1()
297 reduceBlock(s_energy); in sum_energy_kernel1()
312 reduceBlock(s_energy); in sum_energy_kernel2()
324 reduceBlock(s_energy); in sum_energy_kernel2()
666 reduceBlock(dipole); in slabcorr_energy_kernel()
H A Dcrm_cuda_utils.cu297 static __device__ inline void reduceBlock(float* data) in reduceBlock() function
317 static __device__ inline void reduceBlock(int* data) in reduceBlock() function
337 static __device__ inline void reduceBlock(unsigned int* data) in reduceBlock() function
357 static __device__ inline void reduceBlock(double* data) in reduceBlock() function
H A Dpair_tersoff_cuda_kernel_nc.cu34 reduceBlock(shared); in PairVirialCompute_A_Kernel_Template()
39 reduceBlock(shared + 0 * blockDim.x); in PairVirialCompute_A_Kernel_Template()
40 reduceBlock(shared + 1 * blockDim.x); in PairVirialCompute_A_Kernel_Template()
41 reduceBlock(shared + 2 * blockDim.x); in PairVirialCompute_A_Kernel_Template()
42 reduceBlock(shared + 3 * blockDim.x); in PairVirialCompute_A_Kernel_Template()
43 reduceBlock(shared + 4 * blockDim.x); in PairVirialCompute_A_Kernel_Template()
44 reduceBlock(shared + 5 * blockDim.x); in PairVirialCompute_A_Kernel_Template()
H A Dpair_eam_cuda_kernel_nc.cu150 reduceBlock(sharedmem); in PairEAMCuda_Kernel1()
/dports/graphics/opencv/opencv-4.5.3/contrib/modules/cudev/include/opencv2/cudev/block/
H A Dvec_distance.hpp80 … template <int THREAD_DIM> __device__ __forceinline__ void reduceBlock(result_type* smem, uint tid) in reduceBlock() function
109 … template <int THREAD_DIM> __device__ __forceinline__ void reduceBlock(result_type* smem, uint tid) in reduceBlock() function
142 … template <int THREAD_DIM> __device__ __forceinline__ void reduceBlock(result_type* smem, uint tid) in reduceBlock() function
174 … template <int THREAD_DIM> __device__ __forceinline__ void reduceBlock(result_type* smem, uint tid) in reduceBlock() function
/dports/devel/llvm-cheri/llvm-project-37c49ff00e3eadce5d8703fdc4497f28458c64a8/mlir/lib/Conversion/SCFToStandard/
H A DSCFToStandard.cpp382 Block &reduceBlock = reduce.reductionOperator().front(); in matchAndRewrite() local
383 mapping.map(reduceBlock.getArgument(0), mapping.lookupOrDefault(arg)); in matchAndRewrite()
384 mapping.map(reduceBlock.getArgument(1), in matchAndRewrite()
386 for (auto &nested : reduceBlock.without_terminator()) in matchAndRewrite()
389 mapping.lookup(reduceBlock.getTerminator()->getOperand(0))); in matchAndRewrite()
/dports/www/chromium-legacy/chromium-88.0.4324.182/third_party/llvm/mlir/lib/Conversion/SCFToStandard/
H A DSCFToStandard.cpp459 Block &reduceBlock = reduce.reductionOperator().front(); in matchAndRewrite() local
460 mapping.map(reduceBlock.getArgument(0), mapping.lookupOrDefault(arg)); in matchAndRewrite()
461 mapping.map(reduceBlock.getArgument(1), in matchAndRewrite()
463 for (auto &nested : reduceBlock.without_terminator()) in matchAndRewrite()
466 mapping.lookup(reduceBlock.getTerminator()->getOperand(0))); in matchAndRewrite()
/dports/devel/wasi-compiler-rt12/llvm-project-12.0.1.src/mlir/lib/Conversion/SCFToStandard/
H A DSCFToStandard.cpp451 Block &reduceBlock = reduce.reductionOperator().front(); in matchAndRewrite() local
453 yieldOperands.push_back(reduceBlock.getTerminator()->getOperand(0)); in matchAndRewrite()
454 rewriter.eraseOp(reduceBlock.getTerminator()); in matchAndRewrite()
455 rewriter.mergeBlockBefore(&reduceBlock, &op, {arg, reduce.operand()}); in matchAndRewrite()
/dports/devel/llvm12/llvm-project-12.0.1.src/mlir/lib/Conversion/SCFToStandard/
H A DSCFToStandard.cpp451 Block &reduceBlock = reduce.reductionOperator().front(); in matchAndRewrite() local
453 yieldOperands.push_back(reduceBlock.getTerminator()->getOperand(0)); in matchAndRewrite()
454 rewriter.eraseOp(reduceBlock.getTerminator()); in matchAndRewrite()
455 rewriter.mergeBlockBefore(&reduceBlock, &op, {arg, reduce.operand()}); in matchAndRewrite()
/dports/devel/llvm-devel/llvm-project-f05c95f10fc1d8171071735af8ad3a9e87633120/mlir/lib/Conversion/SCFToStandard/
H A DSCFToStandard.cpp490 Block &reduceBlock = reduce.reductionOperator().front(); in matchAndRewrite() local
492 yieldOperands.push_back(reduceBlock.getTerminator()->getOperand(0)); in matchAndRewrite()
493 rewriter.eraseOp(reduceBlock.getTerminator()); in matchAndRewrite()
494 rewriter.mergeBlockBefore(&reduceBlock, &op, {arg, reduce.operand()}); in matchAndRewrite()
/dports/devel/wasi-compiler-rt13/llvm-project-13.0.1.src/mlir/lib/Conversion/SCFToStandard/
H A DSCFToStandard.cpp490 Block &reduceBlock = reduce.reductionOperator().front(); in matchAndRewrite() local
492 yieldOperands.push_back(reduceBlock.getTerminator()->getOperand(0)); in matchAndRewrite()
493 rewriter.eraseOp(reduceBlock.getTerminator()); in matchAndRewrite()
494 rewriter.mergeBlockBefore(&reduceBlock, &op, {arg, reduce.operand()}); in matchAndRewrite()
/dports/devel/wasi-libcxx/llvm-project-13.0.1.src/mlir/lib/Conversion/SCFToStandard/
H A DSCFToStandard.cpp490 Block &reduceBlock = reduce.reductionOperator().front(); in matchAndRewrite() local
492 yieldOperands.push_back(reduceBlock.getTerminator()->getOperand(0)); in matchAndRewrite()
493 rewriter.eraseOp(reduceBlock.getTerminator()); in matchAndRewrite()
494 rewriter.mergeBlockBefore(&reduceBlock, &op, {arg, reduce.operand()}); in matchAndRewrite()
/dports/devel/llvm13/llvm-project-13.0.1.src/mlir/lib/Conversion/SCFToStandard/
H A DSCFToStandard.cpp490 Block &reduceBlock = reduce.reductionOperator().front();
492 yieldOperands.push_back(reduceBlock.getTerminator()->getOperand(0));
493 rewriter.eraseOp(reduceBlock.getTerminator()); in add_slave()
494 rewriter.mergeBlockBefore(&reduceBlock, &op, {arg, reduce.operand()});
/dports/devel/llvm-devel/llvm-project-f05c95f10fc1d8171071735af8ad3a9e87633120/mlir/lib/Dialect/SCF/
H A DSCF.cpp1849 Block &reduceBlock = reduce.reductionOperator().front(); in matchAndRewrite() local
1851 mapping.map(reduceBlock.getArgument(0), op.initVals()[initValIndex]); in matchAndRewrite()
1852 mapping.map(reduceBlock.getArgument(1), in matchAndRewrite()
1854 for (auto &reduceBodyOp : reduceBlock.without_terminator()) in matchAndRewrite()
1858 cast<ReduceReturnOp>(reduceBlock.getTerminator()).result()); in matchAndRewrite()
/dports/devel/wasi-compiler-rt13/llvm-project-13.0.1.src/mlir/lib/Dialect/SCF/
H A DSCF.cpp1852 Block &reduceBlock = reduce.reductionOperator().front(); in matchAndRewrite() local
1854 mapping.map(reduceBlock.getArgument(0), op.initVals()[initValIndex]); in matchAndRewrite()
1855 mapping.map(reduceBlock.getArgument(1), in matchAndRewrite()
1857 for (auto &reduceBodyOp : reduceBlock.without_terminator()) in matchAndRewrite()
1861 cast<ReduceReturnOp>(reduceBlock.getTerminator()).result()); in matchAndRewrite()
/dports/devel/wasi-libcxx/llvm-project-13.0.1.src/mlir/lib/Dialect/SCF/
H A DSCF.cpp1852 Block &reduceBlock = reduce.reductionOperator().front(); in matchAndRewrite() local
1854 mapping.map(reduceBlock.getArgument(0), op.initVals()[initValIndex]); in matchAndRewrite()
1855 mapping.map(reduceBlock.getArgument(1), in matchAndRewrite()
1857 for (auto &reduceBodyOp : reduceBlock.without_terminator()) in matchAndRewrite()
1861 cast<ReduceReturnOp>(reduceBlock.getTerminator()).result()); in matchAndRewrite()
/dports/devel/llvm13/llvm-project-13.0.1.src/mlir/lib/Dialect/SCF/
H A DSCF.cpp1852 Block &reduceBlock = reduce.reductionOperator().front(); in matchAndRewrite() local
1854 mapping.map(reduceBlock.getArgument(0), op.initVals()[initValIndex]); in matchAndRewrite()
1855 mapping.map(reduceBlock.getArgument(1), in matchAndRewrite()
1857 for (auto &reduceBodyOp : reduceBlock.without_terminator()) in matchAndRewrite()
1861 cast<ReduceReturnOp>(reduceBlock.getTerminator()).result()); in matchAndRewrite()