Lines Matching refs:basis_data
148 CeedBasis_Cuda_shared *basis_data; in CeedOperatorBuildKernelData_Cuda_gen() local
152 CeedCallBackend(CeedBasisGetData(basis, &basis_data)); in CeedOperatorBuildKernelData_Cuda_gen()
153 *use_3d_slices = basis_data->d_collo_grad_1d && (was_grad_found ? *use_3d_slices : true); in CeedOperatorBuildKernelData_Cuda_gen()
163 CeedBasis_Cuda_shared *basis_data; in CeedOperatorBuildKernelData_Cuda_gen() local
167 CeedCallBackend(CeedBasisGetData(basis, &basis_data)); in CeedOperatorBuildKernelData_Cuda_gen()
168 *use_3d_slices = basis_data->d_collo_grad_1d && (was_grad_found ? *use_3d_slices : true); in CeedOperatorBuildKernelData_Cuda_gen()
204 CeedBasis_Cuda_shared *basis_data; in CeedOperatorBuildKernelFieldData_Cuda_gen() local
220 CeedCallBackend(CeedBasisGetData(basis, &basis_data)); in CeedOperatorBuildKernelFieldData_Cuda_gen()
248 if (!basis_data->d_chebyshev_interp_1d) { in CeedOperatorBuildKernelFieldData_Cuda_gen()
255 …CeedCallCuda(CeedBasisReturnCeed(basis), cudaMalloc((void **)&basis_data->d_chebyshev_interp_1d, i… in CeedOperatorBuildKernelFieldData_Cuda_gen()
257 …cudaMemcpy(basis_data->d_chebyshev_interp_1d, chebyshev_interp_1d, interp_bytes, cudaMemcpyHostToD… in CeedOperatorBuildKernelFieldData_Cuda_gen()
260 if (is_input) data->B.inputs[i] = basis_data->d_chebyshev_interp_1d; in CeedOperatorBuildKernelFieldData_Cuda_gen()
261 else data->B.outputs[i] = basis_data->d_chebyshev_interp_1d; in CeedOperatorBuildKernelFieldData_Cuda_gen()
264 if (is_input) data->B.inputs[i] = basis_data->d_interp_1d; in CeedOperatorBuildKernelFieldData_Cuda_gen()
265 else data->B.outputs[i] = basis_data->d_interp_1d; in CeedOperatorBuildKernelFieldData_Cuda_gen()
286 if (!basis_data->d_chebyshev_interp_1d) { in CeedOperatorBuildKernelFieldData_Cuda_gen()
293 …CeedCallCuda(CeedBasisReturnCeed(basis), cudaMalloc((void **)&basis_data->d_chebyshev_interp_1d, i… in CeedOperatorBuildKernelFieldData_Cuda_gen()
295 …cudaMemcpy(basis_data->d_chebyshev_interp_1d, chebyshev_interp_1d, interp_bytes, cudaMemcpyHostToD… in CeedOperatorBuildKernelFieldData_Cuda_gen()
298 if (is_input) data->B.inputs[i] = basis_data->d_chebyshev_interp_1d; in CeedOperatorBuildKernelFieldData_Cuda_gen()
299 else data->B.outputs[i] = basis_data->d_chebyshev_interp_1d; in CeedOperatorBuildKernelFieldData_Cuda_gen()
302 if (is_input) data->B.inputs[i] = basis_data->d_interp_1d; in CeedOperatorBuildKernelFieldData_Cuda_gen()
303 else data->B.outputs[i] = basis_data->d_interp_1d; in CeedOperatorBuildKernelFieldData_Cuda_gen()
324 if (is_input) data->G.inputs[i] = basis_data->d_collo_grad_1d; in CeedOperatorBuildKernelFieldData_Cuda_gen()
325 else data->G.outputs[i] = basis_data->d_collo_grad_1d; in CeedOperatorBuildKernelFieldData_Cuda_gen()
337 bool has_collo_grad = basis_data->d_collo_grad_1d; in CeedOperatorBuildKernelFieldData_Cuda_gen()
339 …if (is_input) data->G.inputs[i] = has_collo_grad ? basis_data->d_collo_grad_1d : basis_data->d_gra… in CeedOperatorBuildKernelFieldData_Cuda_gen()
340 … else data->G.outputs[i] = has_collo_grad ? basis_data->d_collo_grad_1d : basis_data->d_grad_1d; in CeedOperatorBuildKernelFieldData_Cuda_gen()
613 CeedBasis_Cuda_shared *basis_data; in CeedOperatorBuildKernelBasis_Cuda_gen() local
619 CeedCallBackend(CeedBasisGetData(basis, &basis_data)); in CeedOperatorBuildKernelBasis_Cuda_gen()
620 data->W = basis_data->d_q_weight_1d; in CeedOperatorBuildKernelBasis_Cuda_gen()