Skip to content
New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Fix prehash crashing on newer NVIDIA drivers #54

Open
wants to merge 1 commit into
base: main
Choose a base branch
from
Open
Changes from all commits
Commits
File filter

Filter by extension

Filter by extension

Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
Fix prehash crashing on newer NVIDIA drivers
Using any NVIDIA driver that is part of a 5XX release branch (such as any drivers released after and including 510.39.01 for Linux or 511.23 for Windows) will cause the miner to crash with cudaErrorIllegalAddress (700) at the prehash stage. Downgrading to an older driver such as 470.141.03 has been a stop-gap solution to getting the miner running again without this patch.

The issue is coming from this line: https://github.com/mhssamadani/Autolykos2_NV_Miner/blob/4de5cea0952a86d5d287d9788146e0840cb9b61d/secp256k1/src/prehash.cu#L251

Simply signaling to the compiler to disable unrolling for this one loop seems to be enough to make the driver happy with no measurable performance difference as far as I have measured.

Fixes #51 #53
Timbers007 authored Oct 15, 2022

Verified

This commit was created on GitHub.com and signed with GitHub’s verified signature. The key has expired.
commit 36cdbc0e5736064da9b9623dc8f22e0a1fc07aa9
2 changes: 1 addition & 1 deletion secp256k1/src/prehash.cu
Original file line number Diff line number Diff line change
@@ -247,7 +247,7 @@ __global__ void InitPrehash(
// Dump result to global memory -- BIG ENDIAN
//====================================================================//

#pragma unroll
#pragma unroll 1
for (int i = 0; i < 4; ++i) store64(&(((uint64_t *)hashes)[(tid + 1) * 4 - i - 1]), h[i]);
((uint8_t *)hashes)[tid * 32 + 31] = 0;
}