From 00734f1bdde59f8fbe2747c872b0da18a3ea8e4e Mon Sep 17 00:00:00 2001 From: Bernhard Manfred Gruber Date: Mon, 27 Feb 2023 16:43:48 +0100 Subject: [PATCH] Add `View::blobs()` accessor Then rename the `storageBlobs` data member to `m_blobs` and make it private. --- docs/pages/mappings.rst | 4 +-- examples/alpaka/asyncblur/asyncblur.cpp | 10 ++----- examples/alpaka/daxpy/daxpy.cpp | 12 ++++---- examples/alpaka/nbody/nbody.cpp | 4 +-- examples/alpaka/pic/pic.cpp | 6 ++-- examples/alpaka/vectoradd/vectoradd.cpp | 8 ++--- examples/bufferguard/bufferguard.cpp | 2 +- examples/cuda/nbody/nbody.cu | 20 ++++++------- examples/nbody/nbody.cpp | 4 +-- examples/root/lhcb_analysis/lhcb.cpp | 12 ++++---- examples/vectoradd/vectoradd.cpp | 2 +- include/llama/Copy.hpp | 20 +++++-------- include/llama/DumpMapping.hpp | 4 +-- include/llama/View.hpp | 39 ++++++++++++++++--------- tests/common.hpp | 6 ++-- tests/mapping.Bytesplit.cpp | 2 +- tests/mapping.Byteswap.cpp | 2 +- tests/mapping.HeatmapTrace.cpp | 22 +++++++------- tests/view.cpp | 6 ++-- 19 files changed, 94 insertions(+), 91 deletions(-) diff --git a/docs/pages/mappings.rst b/docs/pages/mappings.rst index b011e25e14..9d5d5f7306 100644 --- a/docs/pages/mappings.rst +++ b/docs/pages/mappings.rst @@ -199,7 +199,7 @@ A script for gnuplot visualizing the heatmap can be extracted. auto anyMapping = ...; llama::mapping::Heatmap mapping{anyMapping}; ... - mapping.writeGnuplotDataFileBinary(view.storageBlobs, std::ofstream{"heatmap.data", std::ios::binary}); + mapping.writeGnuplotDataFileBinary(view.blobs(), std::ofstream{"heatmap.data", std::ios::binary}); std::ofstream{"plot.sh"} << mapping.gnuplotScriptBinary; @@ -215,7 +215,7 @@ The mapping adds an additional blob to the blobs of the inner mapping used as st auto anyMapping = ...; llama::mapping::FieldAccessCount mapping{anyMapping}; ... - mapping.printFieldHits(view.storageBlobs); // print report with read and writes to each field + mapping.printFieldHits(view.blobs()); // print report with read and writes to each field The FieldAccessCount mapping uses proxy references to instrument reads and writes. If this is problematic, it can also be configured to return raw C++ references. diff --git a/examples/alpaka/asyncblur/asyncblur.cpp b/examples/alpaka/asyncblur/asyncblur.cpp index 30e7ea7871..acef7fa867 100644 --- a/examples/alpaka/asyncblur/asyncblur.cpp +++ b/examples/alpaka/asyncblur/asyncblur.cpp @@ -330,10 +330,7 @@ try hostChunkView[chunkNr](y, x) = subViewHost(y, x); } for(std::size_t i = 0; i < devMapping.blobCount; i++) - alpaka::memcpy( - queue[chunkNr], - devOldView[chunkNr].storageBlobs[i], - hostChunkView[chunkNr].storageBlobs[i]); + alpaka::memcpy(queue[chunkNr], devOldView[chunkNr].blobs()[i], hostChunkView[chunkNr].blobs()[i]); alpaka::exec( queue[chunkNr], @@ -343,10 +340,7 @@ try llama::shallowCopy(devNewView[chunkNr])); for(std::size_t i = 0; i < devMapping.blobCount; i++) - alpaka::memcpy( - queue[chunkNr], - hostChunkView[chunkNr].storageBlobs[i], - devNewView[chunkNr].storageBlobs[i]); + alpaka::memcpy(queue[chunkNr], hostChunkView[chunkNr].blobs()[i], devNewView[chunkNr].blobs()[i]); } // Wait for not finished tasks on accelerator diff --git a/examples/alpaka/daxpy/daxpy.cpp b/examples/alpaka/daxpy/daxpy.cpp index 6005c09413..cc69b2a613 100644 --- a/examples/alpaka/daxpy/daxpy.cpp +++ b/examples/alpaka/daxpy/daxpy.cpp @@ -102,10 +102,10 @@ void daxpyAlpakaLlama(std::string mappingName, std::ofstream& plotFile, Mapping for(std::size_t i = 0; i < Mapping::blobCount; i++) { - auto vx = alpaka::createView(devHost, &x.storageBlobs[0][0], mapping.blobSize(i)); - auto vy = alpaka::createView(devHost, &y.storageBlobs[0][0], mapping.blobSize(i)); - alpaka::memcpy(queue, viewX.storageBlobs[i], vx); - alpaka::memcpy(queue, viewY.storageBlobs[i], vy); + auto vx = alpaka::createView(devHost, &x.blobs()[0][0], mapping.blobSize(i)); + auto vy = alpaka::createView(devHost, &y.blobs()[0][0], mapping.blobSize(i)); + alpaka::memcpy(queue, viewX.blobs()[i], vx); + alpaka::memcpy(queue, viewY.blobs()[i], vy); } watch.printAndReset("copy H->D"); @@ -145,8 +145,8 @@ void daxpyAlpakaLlama(std::string mappingName, std::ofstream& plotFile, Mapping for(std::size_t i = 0; i < Mapping::blobCount; i++) { - auto vz = alpaka::createView(devHost, &z.storageBlobs[0][0], mapping.blobSize(i)); - alpaka::memcpy(queue, vz, viewZ.storageBlobs[i]); + auto vz = alpaka::createView(devHost, &z.blobs()[0][0], mapping.blobSize(i)); + alpaka::memcpy(queue, vz, viewZ.blobs()[i]); } watch.printAndReset("copy D->H"); diff --git a/examples/alpaka/nbody/nbody.cpp b/examples/alpaka/nbody/nbody.cpp index 119883f642..6cbe2088f5 100644 --- a/examples/alpaka/nbody/nbody.cpp +++ b/examples/alpaka/nbody/nbody.cpp @@ -281,7 +281,7 @@ void run(std::ostream& plotFile) watch.printAndReset("init"); for(std::size_t i = 0; i < mapping.blobCount; i++) - alpaka::memcpy(queue, accView.storageBlobs[i], hostView.storageBlobs[i]); + alpaka::memcpy(queue, accView.blobs()[i], hostView.blobs()[i]); watch.printAndReset("copy H->D"); const auto workdiv = alpaka::WorkDivMembers{ @@ -307,7 +307,7 @@ void run(std::ostream& plotFile) plotFile << std::quoted(title) << "\t" << sumUpdate / steps << '\t' << sumMove / steps << '\n'; for(std::size_t i = 0; i < mapping.blobCount; i++) - alpaka::memcpy(queue, hostView.storageBlobs[i], accView.storageBlobs[i]); + alpaka::memcpy(queue, hostView.blobs()[i], accView.blobs()[i]); watch.printAndReset("copy D->H"); const auto [x, y, z] = hostView(referenceParticleIndex)(tag::Pos{}); diff --git a/examples/alpaka/pic/pic.cpp b/examples/alpaka/pic/pic.cpp index 17515d880c..ac20d56b69 100644 --- a/examples/alpaka/pic/pic.cpp +++ b/examples/alpaka/pic/pic.cpp @@ -379,7 +379,7 @@ auto setup(Queue& queue, const Dev& dev, const DevHost& /*devHost*/) } // std::shuffle(particlesHost.begin(), particlesHost.end(), engine); for(auto i = 0; i < decltype(particleMapping)::blobCount; i++) - alpaka::memcpy(queue, particles.storageBlobs[i], particlesHost.storageBlobs[i]); + alpaka::memcpy(queue, particles.blobs()[i], particlesHost.blobs()[i]); return std::tuple{e, b, j, particles}; } @@ -837,7 +837,7 @@ void run(std::ostream& plotFile) auto copyBlobs = [&](auto& fieldView) { for(auto i = 0; i < fieldMapping.blobCount; i++) - alpaka::memcpy(queue, hostFieldView.storageBlobs[i], fieldView.storageBlobs[i]); + alpaka::memcpy(queue, hostFieldView.blobs()[i], fieldView.blobs()[i]); }; copyBlobs(E); output(n, "E", hostFieldView); @@ -850,7 +850,7 @@ void run(std::ostream& plotFile) const auto particlesMapping = particles.mapping(); auto hostParticleView = llama::allocViewUninitialized(particlesMapping); for(auto i = 0; i < particlesMapping.blobCount; i++) - alpaka::memcpy(queue, hostParticleView.storageBlobs[i], particles.storageBlobs[i]); + alpaka::memcpy(queue, hostParticleView.blobs()[i], particles.blobs()[i]); output(n, hostParticleView); } } diff --git a/examples/alpaka/vectoradd/vectoradd.cpp b/examples/alpaka/vectoradd/vectoradd.cpp index 7db34ee7b1..63736e3c64 100644 --- a/examples/alpaka/vectoradd/vectoradd.cpp +++ b/examples/alpaka/vectoradd/vectoradd.cpp @@ -152,8 +152,8 @@ try const auto blobCount = decltype(mapping)::blobCount; for(std::size_t i = 0; i < blobCount; i++) { - alpaka::memcpy(queue, devA.storageBlobs[i], hostA.storageBlobs[i]); - alpaka::memcpy(queue, devB.storageBlobs[i], hostB.storageBlobs[i]); + alpaka::memcpy(queue, devA.blobs()[i], hostA.blobs()[i]); + alpaka::memcpy(queue, devB.blobs()[i], hostB.blobs()[i]); } chrono.printAndReset("Copy H->D"); @@ -175,8 +175,8 @@ try for(std::size_t i = 0; i < blobCount; i++) { - alpaka::memcpy(queue, hostA.storageBlobs[i], devA.storageBlobs[i]); - alpaka::memcpy(queue, hostB.storageBlobs[i], devB.storageBlobs[i]); + alpaka::memcpy(queue, hostA.blobs()[i], devA.blobs()[i]); + alpaka::memcpy(queue, hostB.blobs()[i], devB.blobs()[i]); } chrono.printAndReset("Copy D->H"); } diff --git a/examples/bufferguard/bufferguard.cpp b/examples/bufferguard/bufferguard.cpp index 5a192f0a3b..91933c8475 100644 --- a/examples/bufferguard/bufferguard.cpp +++ b/examples/bufferguard/bufferguard.cpp @@ -242,7 +242,7 @@ void run(const std::string& mappingName) const auto src = srcBlobs[i]; const auto dst = dstBlobs[i]; assert(mapping.blobSize(src) == mapping.blobSize(dst)); - std::memcpy(&dstView.storageBlobs[dst][0], &srcView.storageBlobs[src][0], mapping.blobSize(src)); + std::memcpy(&dstView.blobs()[dst][0], &srcView.blobs()[src][0], mapping.blobSize(src)); } }; diff --git a/examples/cuda/nbody/nbody.cu b/examples/cuda/nbody/nbody.cu index 271d1a0652..25bd9d8af6 100644 --- a/examples/cuda/nbody/nbody.cu +++ b/examples/cuda/nbody/nbody.cu @@ -247,7 +247,7 @@ try hostView(i) = p; } if constexpr(countFieldAccesses) - hostView.mapping().fieldHits(hostView.storageBlobs) = {}; + hostView.mapping().fieldHits(hostView.blobs()) = {}; watch.printAndReset("init"); @@ -267,18 +267,18 @@ try }; start(); - const auto blobs = hostView.storageBlobs.size() / (heatmap ? 2 : 1); // exclude heatmap blobs + const auto blobs = hostView.blobs().size() / (heatmap ? 2 : 1); // exclude heatmap blobs for(std::size_t i = 0; i < blobs; i++) checkError(cudaMemcpy( - &accView.storageBlobs[i][0], - &hostView.storageBlobs[i][0], + &accView.blobs()[i][0], + &hostView.blobs()[i][0], hostView.mapping().blobSize(i), cudaMemcpyHostToDevice)); if constexpr(heatmap) { auto& hmap = accView.mapping(); for(std::size_t i = 0; i < blobs; i++) - cudaMemsetAsync(hmap.blockHitsPtr(i, accView.storageBlobs), 0, hmap.blockHitsSize(i) * sizeof(CountType)); + cudaMemsetAsync(hmap.blockHitsPtr(i, accView.blobs()), 0, hmap.blockHitsSize(i) * sizeof(CountType)); } std::cout << "copy H->D " << stop() << " s\n"; @@ -309,17 +309,17 @@ try plotFile << std::quoted(title) << "\t" << sumUpdate / steps << '\t' << sumMove / steps << '\n'; start(); - for(std::size_t i = 0; i < hostView.storageBlobs.size(); i++) + for(std::size_t i = 0; i < hostView.blobs().size(); i++) checkError(cudaMemcpy( - &hostView.storageBlobs[i][0], - &accView.storageBlobs[i][0], + &hostView.blobs()[i][0], + &accView.blobs()[i][0], hostView.mapping().blobSize(i), cudaMemcpyDeviceToHost)); std::cout << "copy D->H " << stop() << " s\n"; if constexpr(countFieldAccesses) { - hostView.mapping().printFieldHits(hostView.storageBlobs); + hostView.mapping().printFieldHits(hostView.blobs()); } else if constexpr(heatmap) { @@ -329,7 +329,7 @@ try c = '_'; std::ofstream{"plot_heatmap.sh"} << hostView.mapping().gnuplotScript; hostView.mapping().writeGnuplotDataFile( - hostView.storageBlobs, + hostView.blobs(), std::ofstream{"cuda_nbody_heatmap_" + titleCopy + ".dat"}); } diff --git a/examples/nbody/nbody.cpp b/examples/nbody/nbody.cpp index 8febf63f18..54a96d0f59 100644 --- a/examples/nbody/nbody.cpp +++ b/examples/nbody/nbody.cpp @@ -360,7 +360,7 @@ namespace usellama p(tag::Mass{}) = dist(engine) / FP{100}; } if constexpr(countFieldAccesses) - particles.mapping().fieldHits(particles.storageBlobs) = {}; + particles.mapping().fieldHits(particles.blobs()) = {}; watch.printAndReset("init"); double sumUpdate = 0; @@ -393,7 +393,7 @@ namespace usellama if constexpr(heatmap) std::ofstream("nbody_heatmap_" + mappingName(Mapping) + ".sh") << particles.mapping().toGnuplotScript(); if constexpr(countFieldAccesses) - particles.mapping().printFieldHits(particles.storageBlobs); + particles.mapping().printFieldHits(particles.blobs()); return printReferenceParticle(particles(referenceParticleIndex)(tag::Pos{}).load()); } diff --git a/examples/root/lhcb_analysis/lhcb.cpp b/examples/root/lhcb_analysis/lhcb.cpp index 130a5a014a..c70df18d62 100644 --- a/examples/root/lhcb_analysis/lhcb.cpp +++ b/examples/root/lhcb_analysis/lhcb.cpp @@ -419,7 +419,7 @@ namespace { std::filesystem::create_directories(heatmapFile.parent_path()); const auto& m = v.mapping(); - m.writeGnuplotDataFileBinary(v.storageBlobs, std::ofstream{heatmapFile}); + m.writeGnuplotDataFileBinary(v.blobs(), std::ofstream{heatmapFile}); std::ofstream{heatmapFile.parent_path() / "plot.sh"} << View::Mapping::gnuplotScriptBinary; } @@ -428,13 +428,13 @@ namespace { const auto bc = View::Mapping::blobCount; for(int i = bc / 2; i < bc; i++) - std::memset(&v.storageBlobs[i][0], 0, v.mapping().blobSize(i)); + std::memset(&v.blobs()[i][0], 0, v.mapping().blobSize(i)); } template void clearFieldAccessCounts(View& v) { - v.mapping().fieldHits(v.storageBlobs) = {}; + v.mapping().fieldHits(v.blobs()) = {}; } template @@ -470,7 +470,7 @@ namespace const auto& m = v.mapping(); for(std::size_t i = 0; i < View::Mapping::blobCount / 2; i++) { - auto* bh = m.blockHitsPtr(i, v.storageBlobs); + auto* bh = m.blockHitsPtr(i, v.blobs()); auto size = m.blockHitsSize(i); total += std::count_if(bh, bh + size, [](auto c) { return c > 0; }); } @@ -485,7 +485,7 @@ namespace auto [view, conversionTime] = convertRNTupleToLLAMA(inputFile); if constexpr(llama::mapping::isFieldAccessCount) { - view.mapping().printFieldHits(view.storageBlobs); + view.mapping().printFieldHits(view.blobs()); clearFieldAccessCounts(view); } if constexpr(llama::mapping::isHeatmap) @@ -511,7 +511,7 @@ namespace totalAnalysisTime += analysisTime; } if constexpr(llama::mapping::isFieldAccessCount) - view.mapping().printFieldHits(view.storageBlobs); + view.mapping().printFieldHits(view.blobs()); if constexpr(llama::mapping::isHeatmap) saveHeatmap(view, heatmapFolder + "/" + mappingName + "_analysis.bin"); save(hist, mappingName); diff --git a/examples/vectoradd/vectoradd.cpp b/examples/vectoradd/vectoradd.cpp index 813ac4006f..429a44bcab 100644 --- a/examples/vectoradd/vectoradd.cpp +++ b/examples/vectoradd/vectoradd.cpp @@ -119,7 +119,7 @@ namespace usellama } plotFile << "\"" << mappingname << "\"\t" << acc / steps << '\n'; - return static_cast(c.storageBlobs[0][0]); + return static_cast(c.blobs()[0][0]); } } // namespace usellama diff --git a/include/llama/Copy.hpp b/include/llama/Copy.hpp index 30c3963041..d5c621cafe 100644 --- a/include/llama/Copy.hpp +++ b/include/llama/Copy.hpp @@ -63,8 +63,8 @@ namespace llama // TODO(bgruber): this is maybe not the best parallel copying strategy for(std::size_t i = 0; i < Mapping::blobCount; i++) internal::parallelMemcpy( - &dstView.storageBlobs[i][0], - &srcView.storageBlobs[i][0], + &dstView.blobs()[i][0], + &srcView.blobs()[i][0], dstView.mapping().blobSize(i), threadId, threadCount); @@ -169,12 +169,8 @@ namespace llama static constexpr auto dstIsAoSoA = lanesDst != std::numeric_limits::max(); static_assert(srcIsAoSoA || dstIsAoSoA, "At least one of the mappings must be an AoSoA mapping"); - static_assert( - !srcIsAoSoA || std::tuple_size_v == 1, - "Implementation assumes AoSoA with single blob"); - static_assert( - !dstIsAoSoA || std::tuple_size_v == 1, - "Implementation assumes AoSoA with single blob"); + static_assert(!srcIsAoSoA || SrcMapping::blobCount == 1, "Implementation assumes AoSoA with single blob"); + static_assert(!dstIsAoSoA || DstMapping::blobCount == 1, "Implementation assumes AoSoA with single blob"); const auto flatSize = product(dstView.extents()); @@ -207,21 +203,21 @@ namespace llama auto mapSrc = [&](std::size_t flatArrayIndex, auto rc) LLAMA_LAMBDA_INLINE { if constexpr(srcIsAoSoA) - return &srcView.storageBlobs[0][0] + mapAoSoA(flatArrayIndex, rc, lanesSrc); + return &srcView.blobs()[0][0] + mapAoSoA(flatArrayIndex, rc, lanesSrc); else { const auto [blob, off] = mapSoA(flatArrayIndex, rc, isSrcMB); - return &srcView.storageBlobs[blob][off]; + return &srcView.blobs()[blob][off]; } }; auto mapDst = [&](std::size_t flatArrayIndex, auto rc) LLAMA_LAMBDA_INLINE { if constexpr(dstIsAoSoA) - return &dstView.storageBlobs[0][0] + mapAoSoA(flatArrayIndex, rc, lanesDst); + return &dstView.blobs()[0][0] + mapAoSoA(flatArrayIndex, rc, lanesDst); else { const auto [blob, off] = mapSoA(flatArrayIndex, rc, isDstMB); - return &dstView.storageBlobs[blob][off]; + return &dstView.blobs()[blob][off]; } }; diff --git a/include/llama/DumpMapping.hpp b/include/llama/DumpMapping.hpp index 5f3ec137e0..c4a426b4e6 100644 --- a/include/llama/DumpMapping.hpp +++ b/include/llama/DumpMapping.hpp @@ -108,7 +108,7 @@ namespace llama { const auto& mapping = view.mapping(); for(std::size_t i = 0; i < View::Mapping::blobCount; i++) - std::memset(&view.storageBlobs[i][0], pattern, mapping.blobSize(i)); + std::memset(&view.blobs()[i][0], pattern, mapping.blobSize(i)); } template @@ -127,7 +127,7 @@ namespace llama using Type = GetType; // computed values can come from anywhere, so we can only apply heuristics - auto& blobs = view.storageBlobs; + auto& blobs = view.blobs(); auto&& ref = view.mapping().compute(ai, rc, blobs); // try to find the mapped address in one of the blobs diff --git a/include/llama/View.hpp b/include/llama/View.hpp index 036a7a5b5e..77e4e3966a 100644 --- a/include/llama/View.hpp +++ b/include/llama/View.hpp @@ -440,11 +440,16 @@ namespace llama // NOLINTNEXTLINE(cppcoreguidelines-pro-type-member-init,hicpp-member-init) View() = default; + /// Creates a LLAMA View manually. Prefer the allocations functions \ref allocView and \ref + /// allocViewUninitialized if possible. + /// \param mapping The mapping used by the view to map accesses into memory. + /// \param blobs An array of blobs providing storage space for the mapped data. + /// \param accessor The accessor to use when an access is made through this view. LLAMA_FN_HOST_ACC_INLINE - explicit View(Mapping mapping, Array storageBlobs = {}, Accessor accessor = {}) + explicit View(Mapping mapping, Array blobs = {}, Accessor accessor = {}) : Mapping(std::move(mapping)) , Accessor(std::move(accessor)) - , storageBlobs(std::move(storageBlobs)) + , m_blobs(std::move(blobs)) { } @@ -595,7 +600,15 @@ namespace llama return {ArrayIndexRange{extents()}.end(), this}; } - Array storageBlobs; + LLAMA_FN_HOST_ACC_INLINE auto blobs() -> Array& + { + return m_blobs; + } + + LLAMA_FN_HOST_ACC_INLINE auto blobs() const -> const Array& + { + return m_blobs; + } private: template @@ -604,14 +617,16 @@ namespace llama template LLAMA_FN_HOST_ACC_INLINE auto access(ArrayIndex ai, RecordCoord rc = {}) const -> decltype(auto) { - return accessor()(mapToMemory(mapping(), ai, rc, storageBlobs)); + return accessor()(mapToMemory(mapping(), ai, rc, m_blobs)); } template LLAMA_FN_HOST_ACC_INLINE auto access(ArrayIndex ai, RecordCoord rc = {}) -> decltype(auto) { - return accessor()(mapToMemory(mapping(), ai, rc, storageBlobs)); + return accessor()(mapToMemory(mapping(), ai, rc, m_blobs)); } + + Array m_blobs; }; template @@ -624,11 +639,11 @@ namespace llama { template LLAMA_FN_HOST_ACC_INLINE auto makeTransformedBlobArray( - Blobs& storageBlobs, + Blobs& blobs, const TransformBlobFunc& transformBlob, std::integer_sequence) { - return llama::Array{transformBlob(storageBlobs[Is])...}; + return llama::Array{transformBlob(blobs[Is])...}; } } // namespace internal @@ -638,10 +653,8 @@ namespace llama LLAMA_FN_HOST_ACC_INLINE auto transformBlobs(View& view, const TransformBlobFunc& transformBlob) { constexpr auto blobCount = std::decay_t::Mapping::blobCount; - auto blobs = internal::makeTransformedBlobArray( - view.storageBlobs, - transformBlob, - std::make_index_sequence{}); + auto blobs + = internal::makeTransformedBlobArray(view.blobs(), transformBlob, std::make_index_sequence{}); return llama::View{ view.mapping(), std::move(blobs), @@ -678,7 +691,7 @@ namespace llama { return View{ std::move(view.mapping()), - std::move(view.storageBlobs), + std::move(view.blobs()), std::move(newAccessor)}; } @@ -696,7 +709,7 @@ namespace llama return View{ std::move(newMapping), - std::move(view.storageBlobs), + std::move(view.blobs()), std::move(view.accessor())}; } diff --git a/tests/common.hpp b/tests/common.hpp index e9382e03ca..a13d3b2221 100644 --- a/tests/common.hpp +++ b/tests/common.hpp @@ -135,7 +135,7 @@ void iotaStorage(View& view) for(auto i = 0; i < View::Mapping::blobCount; i++) { auto fillFunc = [val = 0]() mutable { return static_cast(val++); }; - std::generate_n(view.storageBlobs[i].storageBlobs.get(), view.mapping().blobSize(i), fillFunc); + std::generate_n(view.blobs()[i].blobs().get(), view.mapping().blobSize(i), fillFunc); } } @@ -282,10 +282,10 @@ struct TriangleAoSWithComputedNormal : llama::mapping::PackedAoS, - llama::Array& storageBlobs) const + llama::Array& blobs) const { auto fetch = [&](llama::NrAndOffset nrAndOffset) -> double - { return *reinterpret_cast(&storageBlobs[nrAndOffset.nr][nrAndOffset.offset]); }; + { return *reinterpret_cast(&blobs[nrAndOffset.nr][nrAndOffset.offset]); }; const auto ax = fetch(Base::template blobNrAndOffset<0, 0>(ai)); const auto ay = fetch(Base::template blobNrAndOffset<0, 1>(ai)); diff --git a/tests/mapping.Bytesplit.cpp b/tests/mapping.Bytesplit.cpp index 487725036a..a4ec025d7e 100644 --- a/tests/mapping.Bytesplit.cpp +++ b/tests/mapping.Bytesplit.cpp @@ -90,7 +90,7 @@ TEST_CASE("mapping.ByteSplit.SoA.verify") // : (b == 3 || b == 7 || b == 11); const auto isNonZero = (b == 0 || b == 4 || b == 8); for(auto i = 0; i < 128; i++) - CHECK(view.storageBlobs[b][i] == (isNonZero ? static_cast(i) : std::byte{0})); + CHECK(view.blobs()[b][i] == (isNonZero ? static_cast(i) : std::byte{0})); } } diff --git a/tests/mapping.Byteswap.cpp b/tests/mapping.Byteswap.cpp index 78f8179c94..07149f7c7e 100644 --- a/tests/mapping.Byteswap.cpp +++ b/tests/mapping.Byteswap.cpp @@ -40,7 +40,7 @@ TEST_CASE("mapping.Byteswap.CheckBytes") view(0)(tag::Y{}) = 0x1234; view(0)(tag::Z{}) = 0x12345678; - std::byte* p = &view.storageBlobs[0][0]; // NOLINT(readability-container-data-pointer) + std::byte* p = &view.blobs()[0][0]; // NOLINT(readability-container-data-pointer) CHECK(*reinterpret_cast(p) == 0x12); CHECK(*reinterpret_cast(p + 2) == 0x3412); CHECK(*reinterpret_cast(p + 4) == 0x78563412); diff --git a/tests/mapping.HeatmapTrace.cpp b/tests/mapping.HeatmapTrace.cpp index db32ffc485..b90e48da82 100644 --- a/tests/mapping.HeatmapTrace.cpp +++ b/tests/mapping.HeatmapTrace.cpp @@ -44,10 +44,10 @@ namespace auto blockHitsSpan(View view, int i) { #ifdef __cpp_lib_span - const auto span = view.mapping().blockHits(i, view.storageBlobs); + const auto span = view.mapping().blockHits(i, view.blobs()); return std::vector(span.begin(), span.end()); #else - const auto* p = view.mapping().blockHitsPtr(i, view.storageBlobs); + const auto* p = view.mapping().blockHitsPtr(i, view.blobs()); const auto size = view.mapping().blockHitsSize(i); return std::vector(p, p + size); #endif @@ -72,7 +72,7 @@ TEST_CASE("Heatmap.simple") 1, 1, 1, 1, 0, 0, 0, 0, 0, 0, 0, 0}); std::stringstream ss; - hmap.writeGnuplotDataFileAscii(view.storageBlobs, ss, false, 24); + hmap.writeGnuplotDataFileAscii(view.blobs(), ss, false, 24); CHECK(ss.str() == "0 0 0 0 1 1 1 1 0 0 0 0 1 1 1 1 0 0 0 0 0 0 0 0\n"); } @@ -93,11 +93,11 @@ TEST_CASE("Heatmap.perInt") CHECK(blockHitsSpan(view, 0) == std::vector{0, 1, 0, 1, 0, 0}); std::stringstream ss; - hmap.writeGnuplotDataFileAscii(view.storageBlobs, ss, false, 6); + hmap.writeGnuplotDataFileAscii(view.blobs(), ss, false, 6); CHECK(ss.str() == "0 1 0 1 0 0\n"); ss = {}; - hmap.writeGnuplotDataFileAscii(view.storageBlobs, ss, true, 6); + hmap.writeGnuplotDataFileAscii(view.blobs(), ss, true, 6); CHECK(ss.str() == "0 1 0 1 0 0\n"); } @@ -121,7 +121,7 @@ TEST_CASE("Heatmap.nbody") { std::ofstream{"plot_heatmap_ascii.sh"} << particles.mapping().gnuplotScriptAscii; std::stringstream ss; - particles.mapping().writeGnuplotDataFileAscii(particles.storageBlobs, ss); + particles.mapping().writeGnuplotDataFileAscii(particles.blobs(), ss); const auto script = ss.str(); std::ofstream{"heatmap." + name + ".dat"} << script; CHECK(script == expectedScript); @@ -131,7 +131,7 @@ TEST_CASE("Heatmap.nbody") std::ofstream{"plot_heatmap_binary.sh"} << particles.mapping().gnuplotScriptBinary; std::vector v; auto bsd = boost::iostreams::back_inserter(v); - particles.mapping().writeGnuplotDataFileBinary(particles.storageBlobs, bsd); + particles.mapping().writeGnuplotDataFileBinary(particles.blobs(), bsd); std::ofstream{"heatmap." + name + ".bin", std::ios::binary}.write( v.data(), static_cast(v.size())); @@ -176,7 +176,7 @@ TEMPLATE_LIST_TEST_CASE("FieldAccessCount.nbody.mem_locs_computed", "", SizeType auto particles = llama::allocView(llama::mapping::FieldAccessCount{ mapping}); // view initialization also writes! updateAndMove(particles); - auto& hits = particles.mapping().fieldHits(particles.storageBlobs); + auto& hits = particles.mapping().fieldHits(particles.blobs()); CHECK(hits.size() == 7); CHECK(hits[0].memLocsComputed == 10400); CHECK(hits[1].memLocsComputed == 10400); @@ -188,7 +188,7 @@ TEMPLATE_LIST_TEST_CASE("FieldAccessCount.nbody.mem_locs_computed", "", SizeType const std::stringstream buffer; std::streambuf* old = std::cout.rdbuf(buffer.rdbuf()); - particles.mapping().printFieldHits(particles.storageBlobs); + particles.mapping().printFieldHits(particles.blobs()); std::cout.rdbuf(old); CHECK(buffer.str() == R"(Field Size Mlocs cmp Pos.X 8 10400 @@ -215,7 +215,7 @@ TEMPLATE_LIST_TEST_CASE("FieldAccessCount.nbody.reads_writes", "", SizeTypes) auto particles = llama::allocView(llama::mapping::FieldAccessCount{ mapping}); // view initialization also writes! updateAndMove(particles); - auto& hits = particles.mapping().fieldHits(particles.storageBlobs); + auto& hits = particles.mapping().fieldHits(particles.blobs()); CHECK(hits.size() == 7); CHECK(hits[0].reads == 10200); CHECK(hits[1].reads == 10200); @@ -234,7 +234,7 @@ TEMPLATE_LIST_TEST_CASE("FieldAccessCount.nbody.reads_writes", "", SizeTypes) const std::stringstream buffer; std::streambuf* old = std::cout.rdbuf(buffer.rdbuf()); - particles.mapping().printFieldHits(particles.storageBlobs); + particles.mapping().printFieldHits(particles.blobs()); std::cout.rdbuf(old); CHECK(buffer.str() == R"(Field Size Reads Writes Pos.X 8 10200 300 diff --git a/tests/view.cpp b/tests/view.cpp index 5d2d46bb49..eb12c60aa6 100644 --- a/tests/view.cpp +++ b/tests/view.cpp @@ -319,7 +319,7 @@ TEST_CASE("view.transformBlobs") auto copy = llama::transformBlobs( view, [](auto& vector) { return std::deque(vector.begin(), vector.end()); }); - STATIC_REQUIRE(std::is_same_v, std::deque>); + STATIC_REQUIRE(std::is_same_v, std::deque>); iotaCheckView(copy); } @@ -337,8 +337,8 @@ TEMPLATE_TEST_CASE( typename std::decay_t::Mapping, typename std::decay_t::Mapping>); // check that blob start address is the same - for(std::size_t i = 0; i < view.storageBlobs.size(); i++) - CHECK(&view.storageBlobs[i][0] == ©.storageBlobs[i][0]); + for(std::size_t i = 0; i < view.blobs().size(); i++) + CHECK(&view.blobs()[i][0] == ©.blobs()[i][0]); }; const auto copy = llama::shallowCopy(view);