Lines Matching refs:impl
39 …::queue &sycl_queue, const SyclModule_t &sycl_module, CeedInt num_elem, const CeedBasis_Sycl *impl, in CeedBasisApplyInterp_Sycl() argument
41 const CeedInt buf_len = impl->buf_len; in CeedBasisApplyInterp_Sycl()
42 const CeedInt op_len = impl->op_len; in CeedBasisApplyInterp_Sycl()
43 const CeedScalar *interp_1d = impl->d_interp_1d; in CeedBasisApplyInterp_Sycl()
47 const CeedInt work_group_size = CeedIntMin(impl->num_qpts, max_work_group_size); in CeedBasisApplyInterp_Sycl()
142 …::queue &sycl_queue, const SyclModule_t &sycl_module, CeedInt num_elem, const CeedBasis_Sycl *impl, in CeedBasisApplyGrad_Sycl() argument
144 const CeedInt buf_len = impl->buf_len; in CeedBasisApplyGrad_Sycl()
145 const CeedInt op_len = impl->op_len; in CeedBasisApplyGrad_Sycl()
146 const CeedScalar *interp_1d = impl->d_interp_1d; in CeedBasisApplyGrad_Sycl()
147 const CeedScalar *grad_1d = impl->d_grad_1d; in CeedBasisApplyGrad_Sycl()
245 …Weight_Sycl(sycl::queue &sycl_queue, CeedInt num_elem, const CeedBasis_Sycl *impl, CeedScalar *w) { in CeedBasisApplyWeight_Sycl() argument
246 const CeedInt dim = impl->dim; in CeedBasisApplyWeight_Sycl()
247 const CeedInt Q_1d = impl->Q_1d; in CeedBasisApplyWeight_Sycl()
248 const CeedScalar *q_weight_1d = impl->d_q_weight_1d; in CeedBasisApplyWeight_Sycl()
277 CeedBasis_Sycl *impl; in CeedBasisApply_Sycl() local
281 CeedCallBackend(CeedBasisGetData(basis, &impl)); in CeedBasisApply_Sycl()
302 …lBackend(CeedBasisApplyInterp_Sycl<true>(data->sycl_queue, *impl->sycl_module, num_elem, impl, d_u… in CeedBasisApply_Sycl()
304 …Backend(CeedBasisApplyInterp_Sycl<false>(data->sycl_queue, *impl->sycl_module, num_elem, impl, d_u… in CeedBasisApply_Sycl()
309 …allBackend(CeedBasisApplyGrad_Sycl<true>(data->sycl_queue, *impl->sycl_module, num_elem, impl, d_u… in CeedBasisApply_Sycl()
311 …llBackend(CeedBasisApplyGrad_Sycl<false>(data->sycl_queue, *impl->sycl_module, num_elem, impl, d_u… in CeedBasisApply_Sycl()
315 …CeedCheck(impl->d_q_weight_1d, ceed, CEED_ERROR_BACKEND, "%s not supported; q_weight_1d not set", … in CeedBasisApply_Sycl()
316 CeedCallBackend(CeedBasisApplyWeight_Sycl(data->sycl_queue, num_elem, impl, d_v)); in CeedBasisApply_Sycl()
338 …cl::queue &sycl_queue, CeedInt num_elem, CeedInt is_transpose, const CeedBasisNonTensor_Sycl *impl, in CeedBasisApplyNonTensorInterp_Sycl() argument
340 const CeedInt num_comp = impl->num_comp; in CeedBasisApplyNonTensorInterp_Sycl()
341 const CeedInt P = is_transpose ? impl->num_qpts : impl->num_nodes; in CeedBasisApplyNonTensorInterp_Sycl()
342 const CeedInt Q = is_transpose ? impl->num_nodes : impl->num_qpts; in CeedBasisApplyNonTensorInterp_Sycl()
343 const CeedInt stride_0 = is_transpose ? 1 : impl->num_nodes; in CeedBasisApplyNonTensorInterp_Sycl()
344 const CeedInt stride_1 = is_transpose ? impl->num_nodes : 1; in CeedBasisApplyNonTensorInterp_Sycl()
351 const CeedScalar *d_B = impl->d_interp; in CeedBasisApplyNonTensorInterp_Sycl()
379 …cl::queue &sycl_queue, CeedInt num_elem, CeedInt is_transpose, const CeedBasisNonTensor_Sycl *impl, in CeedBasisApplyNonTensorGrad_Sycl() argument
381 const CeedInt num_comp = impl->num_comp; in CeedBasisApplyNonTensorGrad_Sycl()
382 const CeedInt P = is_transpose ? impl->num_qpts : impl->num_nodes; in CeedBasisApplyNonTensorGrad_Sycl()
383 const CeedInt Q = is_transpose ? impl->num_nodes : impl->num_qpts; in CeedBasisApplyNonTensorGrad_Sycl()
384 const CeedInt stride_0 = is_transpose ? 1 : impl->num_nodes; in CeedBasisApplyNonTensorGrad_Sycl()
385 const CeedInt stride_1 = is_transpose ? impl->num_nodes : 1; in CeedBasisApplyNonTensorGrad_Sycl()
395 const CeedInt in_dim = is_transpose ? impl->dim : 1; in CeedBasisApplyNonTensorGrad_Sycl()
396 const CeedInt out_dim = is_transpose ? 1 : impl->dim; in CeedBasisApplyNonTensorGrad_Sycl()
397 const CeedScalar *d_G = impl->d_grad; in CeedBasisApplyNonTensorGrad_Sycl()
435 …(sycl::queue &sycl_queue, CeedInt num_elem, const CeedBasisNonTensor_Sycl *impl, CeedScalar *d_V) { in CeedBasisApplyNonTensorWeight_Sycl() argument
436 const CeedInt num_qpts = impl->num_qpts; in CeedBasisApplyNonTensorWeight_Sycl()
437 const CeedScalar *q_weight = impl->d_q_weight; in CeedBasisApplyNonTensorWeight_Sycl()
462 CeedBasisNonTensor_Sycl *impl; in CeedBasisApplyNonTensor_Sycl() local
466 CeedCallBackend(CeedBasisGetData(basis, &impl)); in CeedBasisApplyNonTensor_Sycl()
486 …kend(CeedBasisApplyNonTensorInterp_Sycl(data->sycl_queue, num_elem, is_transpose, impl, d_u, d_v)); in CeedBasisApplyNonTensor_Sycl()
489 …ackend(CeedBasisApplyNonTensorGrad_Sycl(data->sycl_queue, num_elem, is_transpose, impl, d_u, d_v)); in CeedBasisApplyNonTensor_Sycl()
492 …CeedCheck(impl->d_q_weight, ceed, CEED_ERROR_BACKEND, "%s not supported; q_weights not set", CeedE… in CeedBasisApplyNonTensor_Sycl()
493 CeedCallBackend(CeedBasisApplyNonTensorWeight_Sycl(data->sycl_queue, num_elem, impl, d_v)); in CeedBasisApplyNonTensor_Sycl()
518 CeedBasis_Sycl *impl; in CeedBasisDestroy_Sycl() local
519 CeedCallBackend(CeedBasisGetData(basis, &impl)); in CeedBasisDestroy_Sycl()
526 if (impl->d_q_weight_1d) CeedCallSycl(ceed, sycl::free(impl->d_q_weight_1d, data->sycl_context)); in CeedBasisDestroy_Sycl()
527 CeedCallSycl(ceed, sycl::free(impl->d_interp_1d, data->sycl_context)); in CeedBasisDestroy_Sycl()
528 CeedCallSycl(ceed, sycl::free(impl->d_grad_1d, data->sycl_context)); in CeedBasisDestroy_Sycl()
530 CeedCallBackend(CeedFree(&impl)); in CeedBasisDestroy_Sycl()
541 CeedBasisNonTensor_Sycl *impl; in CeedBasisDestroyNonTensor_Sycl() local
542 CeedCallBackend(CeedBasisGetData(basis, &impl)); in CeedBasisDestroyNonTensor_Sycl()
549 if (impl->d_q_weight) CeedCallSycl(ceed, sycl::free(impl->d_q_weight, data->sycl_context)); in CeedBasisDestroyNonTensor_Sycl()
550 CeedCallSycl(ceed, sycl::free(impl->d_interp, data->sycl_context)); in CeedBasisDestroyNonTensor_Sycl()
551 CeedCallSycl(ceed, sycl::free(impl->d_grad, data->sycl_context)); in CeedBasisDestroyNonTensor_Sycl()
553 CeedCallBackend(CeedFree(&impl)); in CeedBasisDestroyNonTensor_Sycl()
564 CeedBasis_Sycl *impl; in CeedBasisCreateTensorH1_Sycl() local
568 CeedCallBackend(CeedCalloc(1, &impl)); in CeedBasisCreateTensorH1_Sycl()
577 impl->dim = dim; in CeedBasisCreateTensorH1_Sycl()
578 impl->P_1d = P_1d; in CeedBasisCreateTensorH1_Sycl()
579 impl->Q_1d = Q_1d; in CeedBasisCreateTensorH1_Sycl()
580 impl->num_comp = num_comp; in CeedBasisCreateTensorH1_Sycl()
581 impl->num_nodes = num_nodes; in CeedBasisCreateTensorH1_Sycl()
582 impl->num_qpts = num_qpts; in CeedBasisCreateTensorH1_Sycl()
583 impl->buf_len = num_comp * CeedIntMax(num_nodes, num_qpts); in CeedBasisCreateTensorH1_Sycl()
584 impl->op_len = Q_1d * P_1d; in CeedBasisCreateTensorH1_Sycl()
592 …CeedCallSycl(ceed, impl->d_q_weight_1d = sycl::malloc_device<CeedScalar>(Q_1d, data->sycl_device, … in CeedBasisCreateTensorH1_Sycl()
593 …sycl::event copy_weight = data->sycl_queue.copy<CeedScalar>(q_weight_1d, impl->d_q_weight_1d, Q_1d… in CeedBasisCreateTensorH1_Sycl()
598 …CeedCallSycl(ceed, impl->d_interp_1d = sycl::malloc_device<CeedScalar>(interp_length, data->sycl_d… in CeedBasisCreateTensorH1_Sycl()
599 …sycl::event copy_interp = data->sycl_queue.copy<CeedScalar>(interp_1d, impl->d_interp_1d, interp_l… in CeedBasisCreateTensorH1_Sycl()
602 …CeedCallSycl(ceed, impl->d_grad_1d = sycl::malloc_device<CeedScalar>(interp_length, data->sycl_dev… in CeedBasisCreateTensorH1_Sycl()
603 …sycl::event copy_grad = data->sycl_queue.copy<CeedScalar>(grad_1d, impl->d_grad_1d, interp_length,… in CeedBasisCreateTensorH1_Sycl()
617 CeedCallSycl(ceed, impl->sycl_module = new SyclModule_t(sycl::build(input_bundle))); in CeedBasisCreateTensorH1_Sycl()
619 CeedCallBackend(CeedBasisSetData(basis, impl)); in CeedBasisCreateTensorH1_Sycl()
634 CeedBasisNonTensor_Sycl *impl; in CeedBasisCreateH1_Sycl() local
638 CeedCallBackend(CeedCalloc(1, &impl)); in CeedBasisCreateH1_Sycl()
644 impl->dim = dim; in CeedBasisCreateH1_Sycl()
645 impl->num_comp = num_comp; in CeedBasisCreateH1_Sycl()
646 impl->num_nodes = num_nodes; in CeedBasisCreateH1_Sycl()
647 impl->num_qpts = num_qpts; in CeedBasisCreateH1_Sycl()
655 …CeedCallSycl(ceed, impl->d_q_weight = sycl::malloc_device<CeedScalar>(num_qpts, data->sycl_device,… in CeedBasisCreateH1_Sycl()
656 …sycl::event copy_weight = data->sycl_queue.copy<CeedScalar>(q_weight, impl->d_q_weight, num_qpts, … in CeedBasisCreateH1_Sycl()
661 …CeedCallSycl(ceed, impl->d_interp = sycl::malloc_device<CeedScalar>(interp_length, data->sycl_devi… in CeedBasisCreateH1_Sycl()
662 …sycl::event copy_interp = data->sycl_queue.copy<CeedScalar>(interp, impl->d_interp, interp_length,… in CeedBasisCreateH1_Sycl()
666 …CeedCallSycl(ceed, impl->d_grad = sycl::malloc_device<CeedScalar>(grad_length, data->sycl_device, … in CeedBasisCreateH1_Sycl()
667 sycl::event copy_grad = data->sycl_queue.copy<CeedScalar>(grad, impl->d_grad, grad_length, e); in CeedBasisCreateH1_Sycl()
672 CeedCallBackend(CeedBasisSetData(basis, impl)); in CeedBasisCreateH1_Sycl()