X-Git-Url: http://biod.pnpi.spb.ru/gitweb/?p=alexxy%2Fgromacs.git;a=blobdiff_plain;f=src%2Fgromacs%2Fmdlib%2Flincs_gpu_internal_sycl.cpp;fp=src%2Fgromacs%2Fmdlib%2Flincs_gpu_internal_sycl.cpp;h=60b10f3f8430a8eab42760e77ae0d8f3c820f55c;hp=658a45b32c7f6fa7ac9aa8c86a09ad4941aaaafe;hb=69d9df0fb1f51869e7f266416c1c3dbe637360ac;hpb=d043b17b33a2e6f4e5cfb8ead27ee9a40b27adb8 diff --git a/src/gromacs/mdlib/lincs_gpu_internal_sycl.cpp b/src/gromacs/mdlib/lincs_gpu_internal_sycl.cpp index 658a45b32c..60b10f3f84 100644 --- a/src/gromacs/mdlib/lincs_gpu_internal_sycl.cpp +++ b/src/gromacs/mdlib/lincs_gpu_internal_sycl.cpp @@ -124,25 +124,25 @@ auto lincsKernel(cl::sycl::handler& cgh, OptionalAccessor a_virialScaled, PbcAiuc pbcAiuc) { - cgh.require(a_constraints); - cgh.require(a_constraintsTargetLengths); + a_constraints.bind(cgh); + a_constraintsTargetLengths.bind(cgh); if constexpr (haveCoupledConstraints) { - cgh.require(a_coupledConstraintsCounts); - cgh.require(a_coupledConstraintsIndices); - cgh.require(a_massFactors); - cgh.require(a_matrixA); + a_coupledConstraintsCounts.bind(cgh); + a_coupledConstraintsIndices.bind(cgh); + a_massFactors.bind(cgh); + a_matrixA.bind(cgh); } - cgh.require(a_inverseMasses); - cgh.require(a_x); - cgh.require(a_xp); + a_inverseMasses.bind(cgh); + a_x.bind(cgh); + a_xp.bind(cgh); if constexpr (updateVelocities) { - cgh.require(a_v); + a_v.bind(cgh); } if constexpr (computeVirial) { - cgh.require(a_virialScaled); + a_virialScaled.bind(cgh); } /* Shared local memory buffer. Corresponds to sh_r, sm_rhs, and sm_threadVirial in CUDA.