[HN Gopher] Failing to reach DDR4 bandwidth
___________________________________________________________________
Failing to reach DDR4 bandwidth
Author : ashvardanian
Score : 79 points
Date : 2022-02-02 15:20 UTC (7 hours ago)
(HTM) web link (unum.cloud)
(TXT) w3m dump (unum.cloud)
| dragontamer wrote:
| GPUs are in fact easier to reach high performance these days IMO.
| Sure, you gotta learn how to stride memory and possibly do it
| across workgroups, but those GPUs really have patterns that
| easily hit their specs (be it FLOPs or Memory Bandwidth)
|
| GPUs are dumb, but that makes them kinda simpler for these kinds
| of simple tests.
|
| CPUs, in particular the Thread ripper Pro, have very complex
| memory hierarchies.
|
| Unless they experiment with NUMA mode (4 or 8 nodes maybe), I
| wouldn't expect much improvement. To max out the memory Bandwidth
| on a CPU, you really gotta understand MESI / cache snooping /
| false sharing, split up your accesses across the chip let's
| (easier to do with NUMA turned on) and code in a NUMA aware
| manner.
|
| ---------
|
| I'm not saying GPUs are always better. But this particular test
| is really bloody simple and ideal for GPUs.
| ashvardanian wrote:
| Yeah, completely agree. Sadly we live in the convergence era.
| Now GPUs are also multi-chip (starting with MI200) and the
| latencies will likely become unpredictable once again :)
| ggreg84 wrote:
| I bet AMD itself is not able to get even close to the 204 GB/s
| bandwidth that they "promise" for these CPUs.
|
| AMD instructions for measuring peak bandwidth are here:
| https://developer.amd.com/spack/stream-benchmark/ (essentially
| does a multi threaded memcpy). The script there will pin 1 thread
| per CCD, and with 8 threads you'll get close to peak bandwidth on
| CPUs with 8 CCDs (adding more threads doesn't help).
|
| I've run it in the same CPUs that AMD uses there, following their
| instructions to the letter, which should be able to achieve 204
| GB/s, but the best I've seen is about 165 GB/s. That's 80% of
| what AMD advertises.
|
| HOWEVER, THIS NUMBER IS A LIE. It is using the AMD compiler with
| a flag called "-fnt-store" to turn all stores into non-temporal
| stores. This is something that no real-world application in
| practice would actually do because it does not make sense doing
| in general.
|
| Without this flag, the peak bandwidth that AMD scripts report is
| about 124 GB/s. That's 60% of what AMD advertises and the best
| that most applications can aim to obtain in practice (the blog
| post author is really close to it).
|
| The fine print (https://www.amd.com/en/products/cpu/amd-
| epyc-7742) does say that 204 GB/s is a "theoretical" number.
|
| So I bought a "theoretical" 204 GB/s CPU that delivers 60% of
| what it "theoretically" promised.
|
| I feel scammed.
|
| ---
|
| Instead, if I run STREAM on my GPU, I get close to 100% of what
| the GPU promises, and about 90% or so of it in practice.
| sliken wrote:
| You seem surprised, this is pretty much the rule for ALL CPUs.
| The 204GB/sec number is simply DDR4-3200 x 64 bits (width of
| one memory channel) x 8 (number of memory channels =
| 204,800MB/sec or "roughly" 204GB/sec (assuming GB = 10^9 not
| 2^30). This is called "peak" bandwidth, as is a never to exceed
| number.
|
| 60-80% of peak is common with tuned code, and less without.
| Sure GPUs, which are all about bandwidth do better. Much of a
| GPU design is about throughput and hiding latency with multiple
| requests/threads/whatever. CPUs have different design goals and
| worry much more about latency and often have to worry about
| NUMA/memory coherency. Not to mention in the general case often
| you have one cache miss per core, which isn't necessarily
| enough to saturate the memory bus.
|
| Try apple M1, other ARMs, Alpha, PA-risc, Sparc, Intel, etc and
| you'll find the same story.
| ashvardanian wrote:
| Those are great links, thanks for sharing! My memcpy-s were
| even slower :)
| dragontamer wrote:
| I'm somewhat surprised that AMD's stream-benchmark uses
| "transparent_hugepage" rather than explicitly using huge-pages.
|
| Has Linux's transparent_hugepage feature gotten good enough to
| use in today's environments? Maybe its good enough for STREAM-
| benchmarking (simple memory patterns) ??
|
| Or is "STREAM" unable to use huge-pages directly (so we require
| the OS to try and guess appropriate usage?) ??
| blihp wrote:
| Yes, it works quite well. Explicit 1gb huge pages also
| provides significant improvements for the right workloads.
| You pretty much have to start taking advantage of features
| like these to get anywhere near the theoretical maximum
| throughput in real world usage even on the higher core count
| desktop chips.
| [deleted]
| tuetuopay wrote:
| You are missing one critical element: NUMA awareness.
|
| Because of the Chiplet design, Threadrippers and EPYC chips are
| basically NUMA systems, like you would with a multi-socket
| motherboard. This has an impact on both memory latency and
| bandwidth.
|
| Redo your test, but pin each thread to a CPU thread, and its
| memory pool to a stick of ram attached directly to the CCX of
| said thread. The bandwidth you are measuring is the Infinity
| Fabric's bandwidth, NOT the DDR4 bandwidth.
|
| You can read more about infinity fabric and its bandwidth here:
| https://en.wikichip.org/wiki/amd/infinity_fabric
| lostmsu wrote:
| I upvoted this knowing the explanation but not feeling myself
| confident enough to actually write it down in the hope that
| someone more authoritative will. Thank you!
| tester756 wrote:
| but it only proves his point
| bravetraveler wrote:
| Hadn't even opened the article yet and I was thinking exactly
| this, 'bet they forgot NUMA'
| vardump wrote:
| Came to say exactly this.
|
| A lot of code is wasting a ton of performance on servers
| because of not being NUMA aware. We should also test high
| throughput/performance code on NUMA systems.
|
| Remember to allocate local memory for your threads -- IOW, be
| NUMA aware. Going through interconnect(s) is slow.
| hardolaf wrote:
| I'm in HFT and a huge part of our job for the FPGA drivers is
| managing the NUMA awareness of our software. And we have to
| constantly educate new devs coming into the low level
| software teams about this. Luckily, the SW teams are starting
| to do this training for us but it's always a major headache
| when there's issues. You can easily lose 50%+ of your
| performance by just failing to set a core affinity without
| even considering anything else.
| jjoonathan wrote:
| How is that generally done? Platform specific libraries?
| Utilities that ask the OS to pin a program to a node?
| magicalhippo wrote:
| Windows has a NUMA API[1] to help with these things. For
| pinning a thread you can for example use the
| SetThreadSelectedCpuSetMasks call[2] from what I can
| gather.
|
| [1]: https://docs.microsoft.com/en-
| us/windows/win32/procthread/nu...
|
| [2]: https://docs.microsoft.com/en-
| us/windows/win32/api/processth...
| pkaye wrote:
| In Linux you can use numactl command to pin a process.
| cyber_kinetist wrote:
| libnuma probably does the job.
|
| https://github.com/numactl/numactl
| ashvardanian wrote:
| On a second thought, are you sure Infinity Fabric would be that
| slow? I would expect it to be compared to L3 bandwidth not RAM.
| Plus, the article you gave suggests 40 GB/s P2P IFOP with
| slower 2.6 GHz RAM. I can't find the right diagram , but as I
| remember, this CPU had 8 chiplets, still allowing for much more
| bandwidth, depending on the topology.
| wmf wrote:
| AFAIK each CCD gets 50 GB/s over Infinity Fabric which adds
| up to 200 GB/s which is effectively the same as memory
| bandwidth. IF should only be a bottleneck if you want to get
| high bandwidth from few CCDs.
| dragontamer wrote:
| > I would expect it to be compared to L3 bandwidth not RAM.
|
| Infinity Fabric is well known to be slower / lower-bandwidth
| than DRAM actually.
|
| Just one of the peculiarities of this architecture.
|
| ------
|
| That's why understanding MESI is important. If Core#10 writes
| something to DRAM: then Core#45 will not be able to write to
| it until Core#10 releases it (aka: reaches the "invalid"
| state of MESI).
|
| In contrast, writing to DRAM that's "not owned" by anybody is
| one-and-done. No need to communicate to other cores or
| chiplets.
|
| ------
|
| No _singular_ core can reach 204GB/s spec of the
| architecture. You only reach that 204GB/s spec if all the
| cores (or chiplets really) are visiting different sections of
| RAM independently.
|
| This happens in practice often: the whole EPYC / Threadripper
| design is really aimed for virtual machines doing independent
| tasks. If you're trying to write high-performance code on
| EPYC / Threadripper, its important to keep this model in your
| brain.
|
| -------
|
| Once you fix this problem, your next problem is probably
| going to be the page-table / virtual memory system. Maybe you
| should be looking up huge-pages in preparation?
| ashvardanian wrote:
| Somehow, similarly bad numbers were achieved with 56 cores
| working with disjoint parts of 512 GB array. I didnt pin
| RAM to cores, though. Maybe thats the missing piece, but
| it's sad if thats the only way to meet the spec. Almost no
| application can do that in modern cloud deployment
| realities.
| lostmsu wrote:
| On the contrary. Having multiple independent services
| running on a single physical machine will likely give you
| full performance. The easiest way is probably just
| splitting the physical machine into the number of
| VMs/pods equal or larger than CCX count.
| dragontamer wrote:
| https://people.freebsd.org/~gallatin/talks/euro2019.pdf
| Here's a talk about a lot of the NUMA-issues brought up
| by Netflix + FreeBSD. Its 200 Gb/s (bits), not bytes, but
| I'd expect you to be running into a lot of the similar
| problems.
|
| Its 200 Gb/s because they're reading from SSDs. But
| still, careful attention to NUMA was the only way they
| could reach full PCIe-bandwidths, which is a similar
| issue to yours (where NUMA is needed to reach full
| Memory/DRAM bandwidth).
|
| --------
|
| Did you set your BIOS into NUMA mode? If you're in
| default BIOS-settings, your RAM is being split up between
| all 4 nodes (Uniform-memory-access mode). At least,
| that's how it was for my Threadripper 1950x (I know
| you've got a totally different model, but who knows?)
|
| I forgot exactly the name of the BIOS setting. But you
| gotta boot your computer up with 4-nodes-per-socket (or
| was it 8-nodes per socket for your computer?) or some
| similar setting. There were some programs on Linux (and
| Windows) that help you see what the OS is seeing.
| kristjansson wrote:
| See also the a more recent talk[0] on progress made on
| the same topic. On EPYC too, so perhaps more applicable
| to OP?
|
| [0]:
| https://people.freebsd.org/~gallatin/talks/euro2021.pdf
| dragontamer wrote:
| The original talk was on 2019 era hardware, for a
| 2019-era talk. Xeon and EPYC are both discussed, albeit
| somewhat dated now but still relevant IMO.
|
| The newer talk is 2021 hardware. IMO, the newer talk
| seems to focus more on the specific details of kTLS and
| PCIe TLS... but there's still stuff in there about NUMA
| for sure.
|
| But yeah, the 2021 talk definitely is on more recent
| hardware: 2021 Xeons and EPYCs.
| kristjansson wrote:
| Ah, dang, that's on me, I forgot the earlier talk
| discussed EPYC chips too.
| bradfa wrote:
| Are you thinking of "NUMA per socket" setting in the
| BIOS?
|
| Dell has a short overview of this focused on Epyc 7000
| series: https://downloads.dell.com/manuals/common/dell-
| emc-dfd-numa-...
| dragontamer wrote:
| That sounds like the one.
|
| Different motherboards call these settings different
| names in their BIOS programs / setup however. So just
| keep an eye out for it.
| dragontamer wrote:
| > Somehow, similarly bad numbers were achieved with 56
| cores working with disjoint parts of 512 GB array
|
| Oh, that's not NUMA at all, now that I'm more carefully
| reading your post. NUMA would involve a "copy" step,
| ensuring that those elements are in NUMA-local memory
| before reading.
|
| Much like how in GPU-programming, you have to worry about
| the physicality of memory, NUMA-aware programming you
| have to memcpy data to the right location before it
| achieves high speeds. Each of the 56-cores needs its
| ~10GBs in "NUMA-local" memory _BEFORE_ you start the
| benchmark.
|
| Yeah, I realize this isn't practical. But... who ever
| said that NUMA use cases are practical? Lol. A lot of
| cases, it makes more sense to just take advantage of
| infinity fabric for simplicity (although its slower, its
| definitely more convenient).
| toast0 wrote:
| NUMA is part of the reality. There's lots of modern
| computing that doesn't acknowledge reality, but that
| doesn't tend to do good things for performance.
|
| Communication isn't cheap, and implicit communication
| makes it hard to tell what's expensive. Reading from a
| pointer could be from L1 cache, or it could be from a
| pci-e card attached to another socket. The code looks the
| same, but the cost is radically different. Organizing how
| your code operates in memory to avoid cross-numa and
| cross-cpu in general isn't always easy, but it's required
| if you want to get the performance you paid for. OTOH,
| smaller nodes get this for free: my dual core desktop
| 'servers' don't have to worry about NUMA and don't have a
| lot of cross-cpu contention either.
| dragontamer wrote:
| > Reading from a pointer could be from L1 cache, or it
| could be from a pci-e card attached to another socket.
|
| The fun one is TLB (translation lookaside buffers) and
| the virtual memory system.
|
| Today's AMD core's have more L3 cache than what the TLB
| can handle with 4k-pages. You need to enable 2MB
| hugepages or 1GB hugepages to even access L3 cache at
| full speeds in practice...
|
| EDIT: Milan-X has 96MB L3 cache per CCX. 4kB-pages would
| require 24,000 (24-thousand) TLB-entries. IIRC, Milan
| only has 2000-TLB-entries. Hurraaahhhhhh....
|
| ------
|
| CPUs are devilishly complicated. It makes optimization
| "fun". Apparently, running "memcpy" requires Ph.D levels
| of study before you can "memcpy" at full speeds these
| days.
| my123 wrote:
| In the same kind of funny subject, GPUs nowadays have
| full MMUs, with TLBs and all present too...
| ashvardanian wrote:
| Thanks, that's a valid suggestion! But in that case I would be
| forced to split my array into smaller chunks in different
| disjoint region of address space, right?
| lmilcin wrote:
| I would add that seeing somebody "forgetting" NUMA makes it
| likely there is a lot of other problems.
|
| NUMA isn't something you "forget" when you try to test
| bandwidth limits. It suggest lack of understanding of how that
| stuff works.
|
| It is as if somebody failed to reach hard drive transfer speeds
| and they "forgot" they need to be looking at the access
| pattern.
| jhokanson wrote:
| Does AMD have folks that you can reach out to regarding this? I
| know Intel has MKL and all the work around its own compiler for
| maximum speed. This seems like it should be trivial for someone
| at AMD to put together as an example of how to do things like
| this correctly ...
| [deleted]
| inetknght wrote:
| The spreadsheet showing a bunch of different approaches:
|
| > Attempt Bandwidth Max Bandwidth Saturation Time to Code
|
| > Parallel STL 87 GB/s 204 GB/s 42.6% 1m
|
| > Best CPU run 122 GB/s 204 GB/s 59.8% 60m
|
| > Thrust 743 GB/s 936 GB/s 79.4% 1m
|
| > Custom CUDA 817 GB/s 936 GB/s 87.3% 30m
|
| > CUB 879 GB/s 936 GB/s 93.9% 5m
|
| Then, benchmarks output:
|
| > 8,115,337,000,378 instructions # 0.18 insn per cycle # 4.71
| stalled cycles per insn (83.33%)
|
| > 1,820,092,697,347 branches # 168.291 M/sec (83.33%)
|
| It's not clear to me which approach represents the output from
| the benchmarks. But that's a _lot_ of branches for the number of
| instructions executed. I suggest that might be a reason that the
| CPU didn't reach high RAM throughput.
| 323 wrote:
| On the other hand, high branch usage is the ideal scenario for
| using CPUs vs GPUs.
| dragontamer wrote:
| Not necessarily.
|
| GPUs are outstandingly good at uniform branches. Only
| divergent branches are GPUs bad at.
| ashvardanian wrote:
| I would further emphasize - locally uniform branches.
| Mostly the cores within the same warp should be well
| synchronized.
| dragontamer wrote:
| It also should be noted that CPUs are _ALSO_ bad at
| divergent branches.
|
| Its just that in the CPU-world, "divergent branches" is
| called "branch misprediction". In the GPU-world, we have
| a better idea of how the "branch predictor" (or the
| equivalent thing to the branch predictor) works. Its
| called SIMD-execution / branch divergence.
|
| -------
|
| That being said: CPUs are way better than GPUs at
| divergent scenarios. Not only is the CPU branch predictor
| able to guess patterns, but its also able to execute in
| parallel to the rest of the CPU (out-of-order speculative
| execution and all).
|
| So in highly divergent branchy code, CPUs work but
| somewhat slowly (if the branch predictor can't predict,
| then its useless). But if there's a simple pattern, GPUs
| are great.
|
| If there's a pattern that can be detected at runtime, but
| is too difficult to program in to a GPU (ex: a binary
| search over a million elements will probably loop roughly
| 20 times), that's where CPUs win exceptionally over GPUs.
|
| The CPU-branch predictor will predict 20-loops in your
| binary search. It might be a little bit wrong (19-loops
| needed or 21-loops needed), but speeding up those
| 20-loops is a huge benefit.
| moonchild wrote:
| Wrt branches, GPUs want spatial locality, and CPUs want
| temporal locality :)
| ashvardanian wrote:
| The numbers are for the last approach, committed to the public
| repo. I agree that it's a lot of branches, but expectedly so.
| Only 0.13% of all branches were missed, so the speculative
| execution almost always works. Even manual 4x unrolling didn't
| help.
| inetknght wrote:
| Indeed? I wonder if you've tried different memory performance
| settings in your BIOS (which "might" invalidate your
| warranty) or different memory modules altogether
| ashvardanian wrote:
| Didn't change in BIOS anything this time. But the entire
| system runs on liquid in a very cold, properly vented room.
| We must have cranked up the BIOS settings during the first
| boot.
| ashvardanian wrote:
| And regarding the unrolling, I also removed the data
| dependency - accumulating into 4 different YMM registers.
| So it's most likely just the Infinity Fabric bottleneck,
| limiting our access to memory.
| jhokanson wrote:
| It is not exactly clear to me what is going on with threads (I
| guess you are using all of them?). I haven't done too much in
| this space but anecdotally I've had better luck if my summation
| is explicitly split into sub-summation tasks. It is not clear if
| that is being done here. It looks like a single summation loop
| that the author is expecting the computer to magically split
| across multiple threads. I'd be interested in seeing what this
| looks like if instead the task were to add chunks of the original
| dataset into results per thread (e.g, first 8000 samples on first
| thread, next 8000 on 2nd thread, etc.), with a final accumulation
| loop across all threads. Again, the author may be trying this and
| this is not my area of expertise but I've had decent luck
| saturating the memory bus with a similar approach.
| ashvardanian wrote:
| Here is the source and the threads: https://github.com/unum-
| cloud/ParallelReductions/blob/fd16d9...
|
| OFC we don't expect the compiler to instantiate them for us,
| it's not OpenMP :) That one we covered in previous articles.
| OpenMP gave us about 50 GB/s with all cores enabled and 80 GB/s
| with part of them disabled.
| a_t48 wrote:
| Is there an advantage to using taskflow for parallel for, if
| you already have another threadpool implementation? I
| recently removed taskflow in a project that was only being
| used for a parallel for loop (as part of a larger refactor,
| the code had a number of issues...), and I'm wondering if
| that was a mistake now that I see that pattern somewhere
| else. :)
| ashvardanian wrote:
| Nope, dont worry :) I did it our of laziness. I didn't want
| to implement a task queue for std::thread-s, so I took
| TaskFlow, as one of the most famous solutions. You can
| definitely get better async task management with enough C++
| experience and time.
| jhokanson wrote:
| My c++ is not great (so it is hard for me to tell what is
| going on) and I'm used to OpenMP where my understanding has
| always been that you tend to get a single thread per
| processor (or per hyper-thread) -- not sure if that is
| guaranteed with the way your code is laid out? Perhaps it
| really is a NUMA issue as others suggest. I will note that
| one other variation I had (as it looks like you are already
| splitting across threads) is that the chunk sizes were
| actually smaller than the # of threads which meant a faster
| thread would take more chunks rather than waiting on the
| slowest thread. Good luck!
| hinkley wrote:
| Doug Lea of Java Memory Model and concurrency note went pretty
| far down this rabbit hole. Not only do you use separate
| counters/queues per thread/core, but you also put empty space
| around them so that you don't accidentally share cache lines. I
| don't know what they do now, but at the time some of the data
| structures in that library used arrays where only every 8th or
| 16th entry is used to avoid two cores trying to read from the
| same cache line.
|
| Typically allocating a separate data structure per actor also
| accomplishes this as a happy accident. If the thread does the
| allocation, then it has a better chance of being in the right
| bank as well.
| ashvardanian wrote:
| Yes, thats needed when you have counters in global memory. In
| that case, instead of just having vector<double> you would
| put each double into a stricture aligned to 64 byte
| addresses. Here all the counters are on local stack, so that
| trick unfortunately wont help
| hinkley wrote:
| For the single threaded version, I believe they have a
| similar problem with auto sums =
| _mm256_set1_ps(0); for (; it + 8 < end; it += 8)
| sums = _mm256_add_ps(_mm256_loadu_ps(it), sums);
|
| Where each SMD op is trying to overwrite to a compact data
| structure.
|
| But in the threaded version https://github.com/unum-
| cloud/ParallelReductions/blob/fd16d9... they have separate
| slots for an accumulator but it's still in a shared vector,
| which most likely has the issue I described.
| StillBored wrote:
| I would try something like this on gravaton2/3 or a recent POWER
| machine, in my experience its a lot easier to get reasonable
| memory perf out of !intel machines.
| bee_rider wrote:
| The threadripper is also a !intel chip.
| ashvardanian wrote:
| Threadripper is by AMD
| bee_rider wrote:
| Yeah that was what I was pointing out -- I was mirroring
| the previous, who I think meant !intel as "not intel."
| tryp wrote:
| Some folks are just not TradeMarksters. They'll teach you
| that Intel makes processors that implement x86 and amd64
| instruction sets. You can Google things on DuckDuckGo.
| They'll blow their nose on the Kleenex that they bought from
| GenericPaperProductsCo. At some restaurants in the US South
| you'll have to let your server know what kind of Coke you
| want: cola, lemon-lime, or the Doctor one.
___________________________________________________________________
(page generated 2022-02-02 23:01 UTC)