Increasing utilization - gdrcopy_copybw
Nobody has claimed this yet.
Assessment
- Difficulty
- 4/5
- Estimated time
- 3-5 days
- Newbie friendliness
- 25/100
- Issue type
- Bug
- Clarity
- Needs clarification
- Activity status
- Stale
- Tech stack
- c, linux
- Domain
- performance
Research direction
Start with the gdrcopy_copybw benchmark and the memcpy_uncached_store_avx implementation mentioned in the issue. Compare the existing _mm256_load_pd and _mm256_stream_pd path with the _mm512 equivalents on the stated NVIDIA A100 80GB PCIe and Gen 4 PCIe setup. Done means documenting the reason for the current width and whether 512-bit operations improve utilization for 32kB-8MB buffers.
Written by the indexing model from the issue text.
Description
Hi,
I am running the gdrcopy_copybw benchmark on NVIDIA A100 80GB PCIe with Gen 4 PCIe.
It appears that the utilization doesn't reach its maximal possible value, getting about 20 GB/s out of the possible 32 GB/s, for buffers of sizes 32kB-8MB.
Upon looking into the code, it appears that in your implementation of memcpy_uncached_store_avx you are using 256 bit functions _mm256_load_pd and _mm256_stream_pd.
What could be the reason for that? Is there a reason for not using 512 bit functions _mm512_load_pd and _mm512_stream_pd instead? Could using the 512 bit functions increase the utilization?
Thanks!
- Dominant language
- C
- Stars
- 1.4k
- Forks
- 198
- PR merge metrics
- No merged PRs in 30d
Getting set up
This project ships no dev container, Dockerfile or contributing guide, so setting up is up to you: start from its README, and see our first-contribution guide for the general steps.
First steps
- Read the whole issue, then the project's contributing guide.
- Comment on the issue to say you are picking it up — it saves two people doing the same work.
- Fork the repository and make your change on a branch.
- Open a pull request that references the issue number.
More from NVIDIA/gdrcopy
-
Difficulty 2/5 1-3 hours Newbie friendliness 78/100
-
Proposal: honor `TMPDIR` in gdrcopy package build scriptsPossibly taken @efegokdemir claimed this 53 days ago. Open
Difficulty 2/5 1-3 hours Newbie friendliness 82/100
-
Difficulty 4/5 3-5 days Newbie friendliness 25/100
-
Difficulty 5/5 Over a week Newbie friendliness 35/100
-
Difficulty 3/5 1-2 days Newbie friendliness 58/100
Similar issues
-
SPIDevice selects an active-high device as soon as it is createdPossibly taken @mikeysklar claimed this today. Open
Difficulty 2/5 1-3 hours Newbie friendliness 73/100
adafruit/circuitpython#11869 ·
Maintainers usually reply within 1 day
-
Difficulty 1/5 Under an hour Newbie friendliness 78/100
gnif/LookingGlass#1352 ·
-
Difficulty 2/5 1-3 hours Newbie friendliness 75/100
Maintainers usually reply within 1 day
-
Difficulty 2/5 1-3 hours Newbie friendliness 85/100
-
Difficulty 2/5 1-3 hours Newbie friendliness 62/100
bazel-contrib/rules_go#4784 · 1 comment ·
Maintainers usually reply within 1 day