| /petsc/src/dm/partitioner/impls/simple/ |
| H A D | partsimple.c | 6 PetscInt gridDim; /* The grid dimension */ member 31 p->gridDim = num; in PetscPartitionerSetFromOptions_Simple() 37 if (p->gridDim < 0) p->gridDim = num; in PetscPartitionerSetFromOptions_Simple() 38 …gridDim == num, PetscObjectComm((PetscObject)part), PETSC_ERR_ARG_INCOMP, "Process grid dimension … in PetscPartitionerSetFromOptions_Simple() 64 for (i = 0; i < p->gridDim; ++i) cells[i] = nodes[i] * procs[i]; in PetscPartitionerPartition_Simple_Grid() 67 for (i = 0; i < p->gridDim; ++i) { in PetscPartitionerPartition_Simple_Grid() 73 for (i = 0; i < p->gridDim; ++i) { in PetscPartitionerPartition_Simple_Grid() 263 p->gridDim = -1; in PetscPartitionerCreate_Simple()
|
| /petsc/src/benchmarks/streams/ |
| H A D | CUDAVersion.cu | 47 idx += blockDim.x * gridDim.x; in set_array() 56 idx += blockDim.x * gridDim.x; in set_array_double() 65 idx += blockDim.x * gridDim.x; in STREAM_Copy() 74 idx += blockDim.x * gridDim.x; in STREAM_Copy_double() 84 if (blockDim.x * gridDim.x < len) return; in STREAM_Copy_Optimized() 95 if (blockDim.x * gridDim.x < len) return; in STREAM_Copy_Optimized_double() 105 idx += blockDim.x * gridDim.x; in STREAM_Scale() 114 idx += blockDim.x * gridDim.x; in STREAM_Scale_double() 124 if (blockDim.x * gridDim.x < len) return; in STREAM_Scale_Optimized() 135 if (blockDim.x * gridDim.x < len) return; in STREAM_Scale_Optimized_double() [all …]
|
| /petsc/src/mat/impls/hypre/ |
| H A D | mhypre_kernels.hpp | 10 PetscInt gridx = gridDim.x * blockDim.x; in ZeroRows() 11 PetscInt gridy = gridDim.y * blockDim.y; in ZeroRows()
|
| /petsc/include/petsc/private/ |
| H A D | cupminterface.hpp | 274 …PETSC_NODISCARD static cudaError_t cupmLaunchKernel(FunctionT &&func, dim3 gridDim, dim3 blockDim,… in PETSC_CUPM_ALIAS_FUNCTION() 279 …nel<util::remove_reference_t<FunctionT>>(std::addressof(func), std::move(gridDim), std::move(block… in PETSC_CUPM_ALIAS_FUNCTION() 430 …PETSC_NODISCARD static hipError_t cupmLaunchKernel(FunctionT &&func, dim3 gridDim, dim3 blockDim, … 434 …return hipLaunchKernel((void *)func, std::move(gridDim), std::move(blockDim), args, sharedMem, std… 844 …PETSC_NODISCARD static cupmError_t cupmLaunchKernel(F &&func, cupmDim3 gridDim, cupmDim3 blockDim,… in cupmLaunchKernel() 846 …ll(util::index_sequence_for<Args...>{}, std::forward<F>(func), std::move(gridDim), std::move(block… in cupmLaunchKernel() 887 …rror_t deduceKernelCall(util::index_sequence<Idx...>, F &&func, cupmDim3 gridDim, cupmDim3 blockDi… in deduceKernelCall() 892 std::move(gridDim), std::move(blockDim), std::move(sharedMem), std::move(stream), in deduceKernelCall()
|
| /petsc/src/sys/objects/device/impls/cupm/ |
| H A D | kernels.hpp | 23 …for (SizeType i = blockIdx.x * blockDim.x + threadIdx.x; i < size; i += blockDim.x * gridDim.x) fu… in grid_stride_1D()
|
| /petsc/src/ksp/pc/impls/pbjacobi/cuda/ |
| H A D | pbjacobi_cuda.cu | 8 const PetscInt gridSize = gridDim.x * blockDim.x; in MatMultBatched()
|
| /petsc/src/mat/impls/aij/mpi/mpihipsparse/ |
| H A D | mpiaijhipsparse.hip.cxx | 123 const PetscCount grid_size = gridDim.x * blockDim.x; in MatPackCOOValues() 130 const PetscCount grid_size = gridDim.x * blockDim.x; in MatAddLocalCOOValues() 147 const PetscCount grid_size = gridDim.x * blockDim.x; in MatAddRemoteCOOValues()
|
| /petsc/src/ksp/pc/impls/vpbjacobi/cuda/ |
| H A D | vpbjacobi_cuda.cu | 83 const PetscInt gridSize = gridDim.x * blockDim.x; in MatMultBatched()
|
| /petsc/src/mat/impls/aij/mpi/mpicusparse/ |
| H A D | mpiaijcusparse.cu | 123 const PetscCount grid_size = gridDim.x * blockDim.x; in MatPackCOOValues() 130 const PetscCount grid_size = gridDim.x * blockDim.x; in MatAddLocalCOOValues() 147 const PetscCount grid_size = gridDim.x * blockDim.x; in MatAddRemoteCOOValues()
|
| /petsc/src/mat/impls/sell/seq/seqhip/ |
| H A D | sellhip.hip.cxx | 196 for (int iter = 0; iter < 1 + (nrows - 1) / (gridDim.x * WARP_SIZE * BLOCKY); iter++) { in matmult_seqsell_tiled_kernel8() 197 …gid = gridDim.x * WARP_SIZE * BLOCKY * iter + blockIdx.x * BLOCKY * WARP_SIZE + threadIdx.y * WARP… in matmult_seqsell_tiled_kernel8() 251 for (int iter = 0; iter < 1 + (nrows - 1) / (gridDim.x * WARP_SIZE * BLOCKY); iter++) { in matmultadd_seqsell_tiled_kernel8() 252 …gid = gridDim.x * WARP_SIZE * BLOCKY * iter + blockIdx.x * BLOCKY * WARP_SIZE + threadIdx.y * WARP… in matmultadd_seqsell_tiled_kernel8()
|
| /petsc/src/mat/impls/sell/seq/seqcuda/ |
| H A D | sellcuda.cu | 192 for (int iter = 0; iter < 1 + (nrows - 1) / (gridDim.x * 32 * BLOCKY); iter++) { in matmult_seqsell_tiled_kernel8() 193 … gid = gridDim.x * 32 * BLOCKY * iter + blockIdx.x * BLOCKY * 32 + threadIdx.y * 32 + threadIdx.x; in matmult_seqsell_tiled_kernel8() 247 for (int iter = 0; iter < 1 + (nrows - 1) / (gridDim.x * 32 * BLOCKY); iter++) { in matmultadd_seqsell_tiled_kernel8() 248 … gid = gridDim.x * 32 * BLOCKY * iter + blockIdx.x * BLOCKY * 32 + threadIdx.y * 32 + threadIdx.x; in matmultadd_seqsell_tiled_kernel8()
|
| /petsc/src/tao/unconstrained/tutorials/ |
| H A D | rosenbrock4.h | 328 PetscInt num_threads = gridDim.x * blockDim.x; in rosenbrock_for_loop()
|
| /petsc/src/vec/vec/impls/seq/cupm/ |
| H A D | vecseqcupm_impl.hpp | 1040 const auto group_entries = (size - 1) / gridDim.x + 1; in EntriesPerGroup() 1059 const auto gdx = gridDim.x; in MDot_kernel()
|
| /petsc/src/mat/impls/aij/seq/seqhipsparse/ |
| H A D | aijhipsparse.hip.cxx | 3773 const PetscCount grid_size = gridDim.x * blockDim.x; in MatAddCOOValues()
|
| /petsc/src/mat/impls/aij/seq/seqcusparse/ |
| H A D | aijcusparse.cu | 4416 const PetscCount grid_size = gridDim.x * blockDim.x; in MatAddCOOValues()
|