+
+namespace gmx::internal
+{
+/*! \brief Helper function to clear device buffer.
+ *
+ * Not applicable to GROMACS's float3 (a.k.a. gmx::RVec) and other custom types.
+ * From SYCL specs: "T must be a scalar value or a SYCL vector type."
+ */
+template<typename ValueType>
+cl::sycl::event fillSyclBufferWithNull(cl::sycl::buffer<ValueType, 1>& buffer,
+ size_t startingOffset,
+ size_t numValues,
+ cl::sycl::queue queue)
+{
+ using cl::sycl::access::mode;
+ const cl::sycl::range<1> range(numValues);
+ const cl::sycl::id<1> offset(startingOffset);
+ const ValueType pattern = ValueType(0); // SYCL vectors support initialization by scalar
+
+ return queue.submit([&](cl::sycl::handler& cgh) {
+ auto d_bufferAccessor =
+ cl::sycl::accessor<ValueType, 1, mode::discard_write>{ buffer, cgh, range, offset };
+ cgh.fill(d_bufferAccessor, pattern);
+ });
+}
+
+//! \brief Helper function to clear device buffer of type float3.
+template<>
+inline cl::sycl::event fillSyclBufferWithNull(cl::sycl::buffer<float3, 1>& buffer,
+ size_t startingOffset,
+ size_t numValues,
+ cl::sycl::queue queue)
+{
+ cl::sycl::buffer<float, 1> bufferAsFloat = buffer.reinterpret<float, 1>(buffer.get_count() * DIM);
+ return fillSyclBufferWithNull<float>(
+ bufferAsFloat, startingOffset * DIM, numValues * DIM, std::move(queue));
+}
+} // namespace gmx::internal
+