23   MatCOOStruct_MPIAIJ *coo = *(MatCOOStruct_MPIAIJ **)data;  in MatCOOStructDestroy_MPIAIJCUSPARSE()  local
26   PetscCall(PetscSFDestroy(&coo->sf));  in MatCOOStructDestroy_MPIAIJCUSPARSE()
27   PetscCallCUDA(cudaFree(coo->Ajmap1));  in MatCOOStructDestroy_MPIAIJCUSPARSE()
28   PetscCallCUDA(cudaFree(coo->Aperm1));  in MatCOOStructDestroy_MPIAIJCUSPARSE()
29   PetscCallCUDA(cudaFree(coo->Bjmap1));  in MatCOOStructDestroy_MPIAIJCUSPARSE()
30   PetscCallCUDA(cudaFree(coo->Bperm1));  in MatCOOStructDestroy_MPIAIJCUSPARSE()
31   PetscCallCUDA(cudaFree(coo->Aimap2));  in MatCOOStructDestroy_MPIAIJCUSPARSE()
32   PetscCallCUDA(cudaFree(coo->Ajmap2));  in MatCOOStructDestroy_MPIAIJCUSPARSE()
33   PetscCallCUDA(cudaFree(coo->Aperm2));  in MatCOOStructDestroy_MPIAIJCUSPARSE()
34   PetscCallCUDA(cudaFree(coo->Bimap2));  in MatCOOStructDestroy_MPIAIJCUSPARSE()
35   PetscCallCUDA(cudaFree(coo->Bjmap2));  in MatCOOStructDestroy_MPIAIJCUSPARSE()
36   PetscCallCUDA(cudaFree(coo->Bperm2));  in MatCOOStructDestroy_MPIAIJCUSPARSE()
37   PetscCallCUDA(cudaFree(coo->Cperm1));  in MatCOOStructDestroy_MPIAIJCUSPARSE()
38   PetscCallCUDA(cudaFree(coo->sendbuf));  in MatCOOStructDestroy_MPIAIJCUSPARSE()
39   PetscCallCUDA(cudaFree(coo->recvbuf));  in MatCOOStructDestroy_MPIAIJCUSPARSE()
40   PetscCall(PetscFree(coo));  in MatCOOStructDestroy_MPIAIJCUSPARSE()
166   MatCOOStruct_MPIAIJ *coo;  in MatSetValuesCOO_MPIAIJCUSPARSE()  local
171   PetscCall(PetscContainerGetPointer(container, &coo));  in MatSetValuesCOO_MPIAIJCUSPARSE()
173   const auto &Annz   = coo->Annz;  in MatSetValuesCOO_MPIAIJCUSPARSE()
174   const auto &Annz2  = coo->Annz2;  in MatSetValuesCOO_MPIAIJCUSPARSE()
175   const auto &Bnnz   = coo->Bnnz;  in MatSetValuesCOO_MPIAIJCUSPARSE()
176   const auto &Bnnz2  = coo->Bnnz2;  in MatSetValuesCOO_MPIAIJCUSPARSE()
177   const auto &vsend  = coo->sendbuf;  in MatSetValuesCOO_MPIAIJCUSPARSE()
178   const auto &v2     = coo->recvbuf;  in MatSetValuesCOO_MPIAIJCUSPARSE()
179   const auto &Ajmap1 = coo->Ajmap1;  in MatSetValuesCOO_MPIAIJCUSPARSE()
180   const auto &Ajmap2 = coo->Ajmap2;  in MatSetValuesCOO_MPIAIJCUSPARSE()
181   const auto &Aimap2 = coo->Aimap2;  in MatSetValuesCOO_MPIAIJCUSPARSE()
182   const auto &Bjmap1 = coo->Bjmap1;  in MatSetValuesCOO_MPIAIJCUSPARSE()
183   const auto &Bjmap2 = coo->Bjmap2;  in MatSetValuesCOO_MPIAIJCUSPARSE()
184   const auto &Bimap2 = coo->Bimap2;  in MatSetValuesCOO_MPIAIJCUSPARSE()
185   const auto &Aperm1 = coo->Aperm1;  in MatSetValuesCOO_MPIAIJCUSPARSE()
186   const auto &Aperm2 = coo->Aperm2;  in MatSetValuesCOO_MPIAIJCUSPARSE()
187   const auto &Bperm1 = coo->Bperm1;  in MatSetValuesCOO_MPIAIJCUSPARSE()
188   const auto &Bperm2 = coo->Bperm2;  in MatSetValuesCOO_MPIAIJCUSPARSE()
189   const auto &Cperm1 = coo->Cperm1;  in MatSetValuesCOO_MPIAIJCUSPARSE()
193     PetscCallCUDA(cudaMalloc((void **)&v1, coo->n * sizeof(PetscScalar)));  in MatSetValuesCOO_MPIAIJCUSPARSE()
194     PetscCallCUDA(cudaMemcpy((void *)v1, v, coo->n * sizeof(PetscScalar), cudaMemcpyHostToDevice));  in MatSetValuesCOO_MPIAIJCUSPARSE()
207   if (coo->sendlen) {  in MatSetValuesCOO_MPIAIJCUSPARSE()
208     MatPackCOOValues<<<(coo->sendlen + 255) / 256, 256>>>(v1, coo->sendlen, Cperm1, vsend);  in MatSetValuesCOO_MPIAIJCUSPARSE()
213 …PetscCall(PetscSFReduceWithMemTypeBegin(coo->sf, MPIU_SCALAR, PETSC_MEMTYPE_CUDA, vsend, PETSC_MEM…  in MatSetValuesCOO_MPIAIJCUSPARSE()
219   PetscCall(PetscSFReduceEnd(coo->sf, MPIU_SCALAR, vsend, v2, MPI_REPLACE));  in MatSetValuesCOO_MPIAIJCUSPARSE()