1 #include <petscdevice.h> 2 #include "ex18.h" 3 4 __global__ void FillValues(PetscInt n, PetscScalar *v) { 5 PetscInt i = blockIdx.x * blockDim.x + threadIdx.x; 6 PetscScalar *s; 7 if (i < n) { 8 s = &v[3 * 3 * i]; 9 for (PetscInt vi = 0; vi < 3; vi++) { 10 for (PetscInt vj = 0; vj < 3; vj++) { s[vi * 3 + vj] = vi + 2 * vj; } 11 } 12 } 13 } 14 15 PetscErrorCode FillMatrixCUDACOO(FEStruct *fe, Mat A) { 16 PetscScalar *v; 17 18 PetscFunctionBeginUser; 19 PetscCallCUDA(cudaMalloc((void **)&v, 3 * 3 * fe->Ne * sizeof(PetscScalar))); 20 FillValues<<<(fe->Ne + 255) / 256, 256>>>(fe->Ne, v); 21 PetscCall(MatSetValuesCOO(A, v, INSERT_VALUES)); 22 PetscCallCUDA(cudaFree(v)); 23 PetscFunctionReturn(0); 24 } 25