[PyCUDA] Long values appear random in Kernel

[Rana, Sanjay]

Hi Everyone,



I am sure it's something quite trivial that I am doing incorrectly but I can't seem to be able to find it.



The following simple pycuda code is printing some random values from the kernel for the long integer array.



Please ignore the arbitrary initialisations in places as this is just for the test.



--code--

import numpy as np

import pycuda.driver as drv

import pycuda.autoinit

from pycuda.compiler import SourceModule



_stats_kernel_source = """

#include <stdio.h>



__global__ void stats_kernel(unsigned long *CROMEID, int numCROMECELLS)

{

// calculate pixel idx

int x = blockIdx.x * blockDim.x + threadIdx.x;

if (x < numCROMECELLS)

{

if (x < 1)

{

unsigned long cid = CROMEID[x];

printf("Value in Kernel at %d is %lu\\n",x,cid);

}

}

}

"""



zonal_stats_cuda = SourceModule(_stats_kernel_source)

zonal_stats_cuda_func = zonal_stats_cuda.get_function("stats_kernel")



numCROMECELLS = np.uint32(91480)

cromeCELLS = np.zeros((numCROMECELLS,), dtype=np.uint64)

long_cromeid = np.uint64(1091194009352)

cromeCELLS[0] = long_cromeid

long_cromeid = np.uint64(1088314008174)

cromeCELLS[1] = long_cromeid

long_cromeid = np.uint64(1091194011015)

cromeCELLS[2] = long_cromeid

print "Num Cells:%d"%(numCROMECELLS)

print "Pre-Kernel Value at 0 in long array is %d"%(cromeCELLS[0])

zonal_stats_cuda_func(drv.In(cromeCELLS), numCROMECELLS, block=(32,1,1))



---output-

Num Cells:91480

Pre-Kernel Value at 0 in long array is 1091194009352

Value in Kernel at 0 is 272316168





Sanjay Rana | Design Directorate

Rural Payments Agency | Room 200 | North Gate House | Reading | RG1 1AF

Jabber: 02077642065 | ext. 42065 | Email: ***@rpa.gsi.gov.uk

Follow us on Twitter @Ruralpay
************************************************************************************
This email and any files transmitted with it are confidential and
intended solely for the use of the individual or entity to whom they
are addressed. If you have received this email in error please notify
the system manager.

This footnote also confirms that this email message has been swept by
Cisco ESA for the presence of computer viruses.
************************************************************************************

[Peter Walsh]

unsigned long = 32bits in cuda.

You probably want unsigned long long = 64bits, which aligns with np.uint64.

Post by Rana, Sanjay
Hi Everyone,
I am sure it’s something quite trivial that I am doing incorrectly but I
can’t seem to be able to find it.
The following simple pycuda code is printing some random values from the
kernel for the long integer array.
Please ignore the arbitrary initialisations in places as this is just for the test.
--code--
import numpy as np
import pycuda.driver as drv
import pycuda.autoinit
from pycuda.compiler import SourceModule
_stats_kernel_source = """
#include <stdio.h>
__global__ void stats_kernel(unsigned long *CROMEID, int numCROMECELLS)
{
// calculate pixel idx
int x = blockIdx.x * blockDim.x + threadIdx.x;
if (x < numCROMECELLS)
{
if (x < 1)
{
unsigned long cid = CROMEID[x];
printf("Value in Kernel at %d is %lu\\n",x,cid);
}
}
}
"""
zonal_stats_cuda = SourceModule(_stats_kernel_source)
zonal_stats_cuda_func = zonal_stats_cuda.get_function("stats_kernel")
numCROMECELLS = np.uint32(91480)
cromeCELLS = np.zeros((numCROMECELLS,), dtype=np.uint64)
long_cromeid = np.uint64(1091194009352)
cromeCELLS[0] = long_cromeid
long_cromeid = np.uint64(1088314008174)
cromeCELLS[1] = long_cromeid
long_cromeid = np.uint64(1091194011015)
cromeCELLS[2] = long_cromeid
print "Num Cells:%d"%(numCROMECELLS)
print "Pre-Kernel Value at 0 in long array is %d"%(cromeCELLS[0])
zonal_stats_cuda_func(drv.In(cromeCELLS), numCROMECELLS, block=(32,1,1))
---output—
Num Cells:91480
Pre-Kernel Value at 0 in long array is 1091194009352
Value in Kernel at 0 is 272316168
Sanjay Rana | Design Directorate
Rural Payments Agency | Room 200 | North Gate House | Reading | RG1 1AF
************************************************************
************************
This email and any files transmitted with it are confidential and
intended solely for the use of the individual or entity to whom they
are addressed. If you have received this email in error please notify
the system manager.
This footnote also confirms that this email message has been swept by
Cisco ESA for the presence of computer viruses.
************************************************************
************************
_______________________________________________
PyCUDA mailing list
https://lists.tiker.net/listinfo/pycuda

[Rana, Sanjay]

Thanks – this worked after a few more tweaks! Revised kernel code is as follow:


__global__ void stats_kernel(unsigned long long *CROMEID, int numCROMECELLS)

{

// calculate pixel idx

int x = blockIdx.x * blockDim.x + threadIdx.x;

if (x < numCROMECELLS)

{

if (x < 1)

{

unsigned long long cid = CROMEID[x];

printf("Value in Kernel at %d is %llu\\n",x,cid);

}

}

}

"""






Sanjay Rana | Design Directorate
Rural Payments Agency | Room 200 | North Gate House | Reading | RG1 1AF
Jabber: 02077642065 | ext. 42065 | Email: ***@rpa.gsi.gov.uk<mailto:***@rpa.gsi.gov.uk>
Follow us on Twitter @Ruralpay

From: Peter Walsh [mailto:***@gmail.com]
Sent: 13 June 2017 09:51
To: Rana, Sanjay
Cc: Andreas Kloeckner; ***@tiker.net
Subject: Re: [PyCUDA] Long values appear random in Kernel

unsigned long = 32bits in cuda.

You probably want unsigned long long = 64bits, which aligns with np.uint64.

On 13 June 2017 at 09:02, Rana, Sanjay <***@rpa.gsi.gov.uk<mailto:***@rpa.gsi.gov.uk>> wrote:

Hi Everyone,



I am sure it’s something quite trivial that I am doing incorrectly but I can’t seem to be able to find it.



The following simple pycuda code is printing some random values from the kernel for the long integer array.



Please ignore the arbitrary initialisations in places as this is just for the test.



--code--

import numpy as np

import pycuda.driver as drv

import pycuda.autoinit

from pycuda.compiler import SourceModule



_stats_kernel_source = """

#include <stdio.h>



__global__ void stats_kernel(unsigned long *CROMEID, int numCROMECELLS)

{

// calculate pixel idx

int x = blockIdx.x * blockDim.x + threadIdx.x;

if (x < numCROMECELLS)

{

if (x < 1)

{

unsigned long cid = CROMEID[x];

printf("Value in Kernel at %d is %lu\\n",x,cid);

}

}

}

"""



zonal_stats_cuda = SourceModule(_stats_kernel_source)

zonal_stats_cuda_func = zonal_stats_cuda.get_function("stats_kernel")



numCROMECELLS = np.uint32(91480)

cromeCELLS = np.zeros((numCROMECELLS,), dtype=np.uint64)

long_cromeid = np.uint64(1091194009352)

cromeCELLS[0] = long_cromeid

long_cromeid = np.uint64(1088314008174)

cromeCELLS[1] = long_cromeid

long_cromeid = np.uint64(1091194011015)

cromeCELLS[2] = long_cromeid

print "Num Cells:%d"%(numCROMECELLS)

print "Pre-Kernel Value at 0 in long array is %d"%(cromeCELLS[0])

zonal_stats_cuda_func(drv.In(cromeCELLS), numCROMECELLS, block=(32,1,1))



---output—

Num Cells:91480

Pre-Kernel Value at 0 in long array is 1091194009352

Value in Kernel at 0 is 272316168





Sanjay Rana | Design Directorate

Rural Payments Agency | Room 200 | North Gate House | Reading | RG1 1AF

Jabber: 02077642065 | ext. 42065 | Email: ***@rpa.gsi.gov.uk<mailto:***@rpa.gsi.gov.uk>

Follow us on Twitter @Ruralpay

************************************************************************************
This email and any files transmitted with it are confidential and
intended solely for the use of the individual or entity to whom they
are addressed. If you have received this email in error please notify
the system manager.

This footnote also confirms that this email message has been swept by
Cisco ESA for the presence of computer viruses.
************************************************************************************

_______________________________________________
PyCUDA mailing list
***@tiker.net<mailto:***@tiker.net>
https://lists.tiker.net/listinfo/pycuda


______________________________________________________________________
This email has been scanned by the Symantec Email Security.cloud service.
For more information please visit http://www.symanteccloud.com
______________________________________________________________________
************************************************************************************
This email and any files transmitted with it are confidential and
intended solely for the use of the individual or entity to whom they
are addressed. If you have received this email in error please notify
the system manager.

This footnote also confirms that this email message has been swept by
Cisco ESA for the presence of computer viruses.
************************************************************************************