modify tests, fix create nonce bytes

Author: camfairchild

__init__.py

@@ -16,4 +16,4 @@
 # OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER 
 # DEALINGS IN THE SOFTWARE.
 
-__version__ = "1.1.0"
+__version__ = "1.1.1"

cubit.pyx

@@ -36,9 +36,9 @@ cdef extern from "kernels/main.hh":
     void reset_cuda_c();
     int runTestSealMeetsDifficulty(unsigned char* seal, unsigned long* limit);
     int runTestLessThan(uint256 a, uint256 b);
-    void runTestCreatePreSeal(unsigned char* pre_seal, uint64 nonce, unsigned char* block_bytes);
-    void runTestCreateNonceBytes(unsigned long long nonce, unsigned char* nonce_bytes);
-    void runTestSealHash(unsigned char* seal, unsigned char* block_hash, uint64 nonce);
+    void runTestCreatePreSeal(unsigned char* pre_seal, uint64 nonce, unsigned char* block_bytes, int dev_id);
+    void runTestCreateNonceBytes(unsigned long long nonce, unsigned char* nonce_bytes, int dev_id);
+    void runTestSealHash(unsigned char* seal, unsigned char* block_hash, uint64 nonce, int dev_id);
     void runTestPreSealHash(unsigned char* seal, unsigned char* preseal_bytes);
     void runTest(unsigned char* data, unsigned long size, unsigned char* digest);
     void runTestKeccak(unsigned char* data, unsigned long size, unsigned char* digest);
@@ -74,12 +74,12 @@ cpdef bytes run_test_keccak(unsigned char* data, unsigned int length):
     finally:
         PyMem_Free(digest_)
 
-cpdef bytes run_test_seal_hash(unsigned char* block_bytes, uint64 nonce):
+cpdef bytes run_test_seal_hash(unsigned char* block_bytes, uint64 nonce, int dev_id):
     cdef unsigned char* digest_ = <unsigned char*> PyMem_Malloc(
         64 * sizeof(unsigned char))
 
     try:
-        runTestSealHash(digest_, block_bytes, nonce)
+        runTestSealHash(digest_, block_bytes, nonce, dev_id)
 
         return digest_[:32]
     finally:
@@ -96,13 +96,13 @@ cpdef bytes run_test_preseal_hash(unsigned char* preseal_bytes):
     finally:
         PyMem_Free(digest_)
 
-cpdef bytes run_test_create_nonce_bytes(uint64 nonce):
+cpdef bytes run_test_create_nonce_bytes(uint64 nonce, int dev_id):
     cdef unsigned char* nonce_bytes = <unsigned char*> PyMem_Malloc(
         8 * sizeof(unsigned char))
     cdef int i
 
     try:
-        runTestCreateNonceBytes(nonce, nonce_bytes)
+        runTestCreateNonceBytes(nonce, nonce_bytes, dev_id)
 
         # Convert digest to python string
         nonce_bytes_str = nonce_bytes
@@ -153,13 +153,13 @@ cpdef int run_test_seal_meets_difficulty(const unsigned char[:] seal, const unsi
         PyMem_Free(seal_)
         PyMem_Free(limit_char)
 
-cpdef bytearray run_test_create_pre_seal(uint64 nonce, unsigned char* block_bytes):
+cpdef bytearray run_test_create_pre_seal(uint64 nonce, unsigned char* block_bytes, int dev_id):
     cdef unsigned char* preseal_bytes = <unsigned char*> PyMem_Malloc(
         40 * sizeof(unsigned char))
     cdef int i
 
     try:
-        runTestCreatePreSeal(preseal_bytes, nonce, block_bytes)
+        runTestCreatePreSeal(preseal_bytes, nonce, block_bytes, dev_id)
 
         return bytearray(preseal_bytes[:40])
     finally:

kernels/main.cu

@@ -85,7 +85,7 @@ __device__ bool seal_meets_difficulty(BYTE* seal, uint256 limit) {
 
 __device__ void create_nonce_bytes(uint64 nonce, BYTE* nonce_bytes) {
     // Convert nonce to bytes (little endian) and store at start of pre_seal;
-    for (int i = 0; i < 5; i++) {
+    for (int i = 0; i < 8; i++) {
         nonce_bytes[i] = (nonce >> (i * 8)) & 0xFF;
     }
 }
@@ -204,6 +204,12 @@ __global__ void test_keccak256(BYTE* data, int size, BYTE* digest) {
 
 __global__ void test_seal_hash(BYTE* seal, BYTE* block_hash, uint64 nonce) {
     create_seal_hash(seal, block_hash, nonce);
+    // print out seal
+    printf("seal: ");
+    for (int i = 0; i < 64; i++) {
+        printf("%02x", seal[i]);
+    }
+    printf("\n");
 }
 
 __global__ void test_preseal_hash(BYTE* seal, BYTE* preseal_bytes) {
@@ -242,7 +248,9 @@ bool runTestSealMeetsDifficulty(BYTE* seal, uint256 limit) {
     return result;
 }
 
-void runTestCreateNonceBytes(uint64 nonce, BYTE* nonce_bytes) {
+void runTestCreateNonceBytes(uint64 nonce, BYTE* nonce_bytes, int dev_id) {
+    checkCudaErrors(cudaSetDevice(dev_id));
+
     BYTE* dev_nonce_bytes = NULL;
     checkCudaErrors(cudaMallocManaged(&dev_nonce_bytes, sizeof(BYTE) * 8));
 
@@ -254,7 +262,9 @@ void runTestCreateNonceBytes(uint64 nonce, BYTE* nonce_bytes) {
     checkCudaErrors(cudaDeviceSynchronize());
 }
 
-void runTestCreatePreSeal(unsigned char* pre_seal, uint64 nonce, unsigned char* block_bytes) {
+void runTestCreatePreSeal(unsigned char* pre_seal, uint64 nonce, unsigned char* block_bytes, int dev_id) {
+    checkCudaErrors(cudaSetDevice(dev_id));
+
     // Test sha256
     BYTE* dev_pre_seal = NULL;
     checkCudaErrors(cudaMallocManaged(&dev_pre_seal, sizeof(BYTE) * 40));
@@ -306,9 +316,11 @@ void runTestKeccak(BYTE* data, unsigned long size, BYTE* digest) {
     checkCudaErrors(cudaDeviceSynchronize());
 }
 
-void runTestSealHash(BYTE* seal, BYTE* block_hash, uint64 nonce) {
+void runTestSealHash(BYTE* seal, BYTE* block_hash, uint64 nonce, int dev_id) {
     BYTE* dev_seal = NULL;
     BYTE* dev_block_hash = NULL;
+    checkCudaErrors(cudaSetDevice(dev_id));
+
     checkCudaErrors(cudaMallocManaged(&dev_seal, 64));
     checkCudaErrors(cudaMallocManaged(&dev_block_hash, 64));
     // Copy data to device

kernels/main.hh

@@ -24,10 +24,10 @@
 extern void reset_cuda_c();
 extern int runTestLessThan(unsigned long* a, unsigned long* b);
 extern bool runTestSealMeetsDifficulty(unsigned char* seal, unsigned long* limit);
-extern void runTestCreatePreSeal(unsigned char* pre_seal, uint64 nonce, unsigned char* block_bytes);
-extern void runTestCreateNonceBytes(unsigned long long nonce, unsigned char* nonce_bytes);
+extern void runTestCreatePreSeal(unsigned char* pre_seal, uint64 nonce, unsigned char* block_bytes, int dev_id);
+extern void runTestCreateNonceBytes(unsigned long long nonce, unsigned char* nonce_bytes, int dev_id);
 extern void runTestPreSealHash(unsigned char* seal, unsigned char* preseal_bytes);
-extern void runTestSealHash(unsigned char* seal, unsigned char* block_hash, uint64 nonce);
+extern void runTestSealHash(unsigned char* seal, unsigned char* block_hash, uint64 nonce, int dev_id);
 extern void runTest(unsigned char* data, unsigned long size, unsigned char* digest);
 extern void runTestKeccak(unsigned char* data, unsigned long size, unsigned char* digest);
 extern uint64 solve_cuda_c(int blockSize, unsigned char* seal, uint64 nonce_start, uint64 update_interval, unsigned long* limit, unsigned char* block_bytes, int dev_id);

pyproject.toml

@@ -6,7 +6,7 @@ build-backend = "setuptools.build_meta"
 
 [project]
 name = "cubit"
-version = "1.1.0"
+version = "1.1.1"
 description = "A python package for CUDA registration on bittensor. "
 readme = "README.md"
 authors = [{ name = "Opentensor Foundation", email = "cameron@opentensor.ai" }]

setup.py

@@ -152,7 +152,7 @@ def build_extensions(self):
     author = 'Opentensor Foundation',
     author_email = 'cameron@opentensor.ai',
     url = 'https://github.com/opentensor/cubit',
-    version = '1.1.0',
+    version = '1.1.1',
 
     ext_modules = cythonize(ext),
 

test.py

@@ -120,8 +120,8 @@ def test_preseal_hash( self ) -> None:
     # Test create nonce bytes from nonce
     def test_create_nonce_bytes( self ) -> bytes:
         print(self._testMethodName)
-        nonce = 100000000
-        nonce_bytes: bytes = run_test_create_nonce_bytes(nonce)
+        nonce = random.randint(int(math.pow(2, 45)), int(math.pow(2, 63)-1))
+        nonce_bytes: bytes = run_test_create_nonce_bytes(nonce, self.dev_id)
         nonce_bytes_2 = self.get_nonce_bytes(nonce)
         # Unhexlify to compare
         nonce_bytes_2 = binascii.unhexlify(nonce_bytes_2)
@@ -131,8 +131,8 @@ def test_create_nonce_bytes( self ) -> bytes:
     # Test create pre seal
     def test_create_pre_seal( self ) -> bytes:
         print(self._testMethodName)
-        nonce = 1304006780
-        pre_seal = run_test_create_pre_seal(nonce, self.block_bytes)
+        nonce = random.randint(int(math.pow(2, 45)), int(math.pow(2, 63)-1))
+        pre_seal = run_test_create_pre_seal(nonce, self.block_bytes, self.dev_id)
 
         nonce_bytes = self.get_nonce_bytes(nonce)
         pre_seal_2 = bytearray(self.hex_bytes_to_u8_list(nonce_bytes + self.block_bytes))
@@ -142,8 +142,8 @@ def test_create_pre_seal( self ) -> bytes:
     # Test block and nonce hash
     def test_seal_hash( self ) -> None:
         print(self._testMethodName)
-        nonce = 0
-        seal = run_test_seal_hash(self.block_bytes, nonce)
+        nonce = random.randint(int(math.pow(2, 45)), int(math.pow(2, 63)-1))
+        seal = run_test_seal_hash(self.block_bytes, nonce, self.dev_id)
         nonce_bytes = self.get_nonce_bytes(nonce)
         pre_seal = nonce_bytes + self.block_bytes