# Handling large integer calculations inside kernel without overflow?

**URL:** <https://forums.developer.nvidia.com/t/handling-large-integer-calculations-inside-kernel-without-overflow/57518>\
**Category:** CUDA Programming and Performance\
**Created:** [January 27, 2018, 5:41pm UTC](https://forums.developer.nvidia.com/t/handling-large-integer-calculations-inside-kernel-without-overflow/57518 "2018-01-27T17:41:18Z")\
**Posts on this page:** 3\
**Page:** 1

<div class="post-metadata">

**Author:** ![nixcoder](https://developer.download.nvidia.com/images/forums/profile-default-devtalk-84.png) [@nixcoder](https://forums.developer.nvidia.com/u/nixcoder)\
**Post date:** [January 27, 2018, 5:41pm UTC](https://forums.developer.nvidia.com/t/handling-large-integer-calculations-inside-kernel-without-overflow/57518/1 "2018-01-27T17:41:18Z")

</div>

Please advise, the best practice to handle large integer calculations inside kernel without overflow. Please suggest the right datatypes that can be used and respective header files required.

Example- Say if I have to multiply 1 billion into 7 billion  
1,000,330,000 x 7,000,567,989

---

<div class="post-metadata">

**Author:** ![striker159](https://developer.download.nvidia.com/images/forums/profile-default-devtalk-84.png) [@striker159](https://forums.developer.nvidia.com/u/striker159)\
**Post date:** [January 27, 2018, 9:12pm UTC](https://forums.developer.nvidia.com/t/handling-large-integer-calculations-inside-kernel-without-overflow/57518/2 "2018-01-27T21:12:40Z")

</div>

What do you mean by large integers? The result of your example fits in a signed 64 bit integer.

```auto
#include <stdio.h>

__global__
void foo(int64_t* c, int64_t a, int64_t b){
	*c = a * b;
}

int main(){
	int64_t* c;
	cudaMallocManaged(&c, sizeof(int64_t));
	foo<<<1,1>>>(c, 1000330000ULL, 7000567989ULL);
	cudaDeviceSynchronize();
	printf("%ld\n", *c);
	cudaFree(c);
}

```

Prints: 7002878176436370000

There are also integer intrinsic functions to calculate the upper 32 / 64 bit of a 32 / 64 bit multiplication  
\_\_mul64hi ( long long int x, long long int y)  
\_\_mulhi (int x, int y)

see [CUDA Math API :: CUDA Toolkit Documentation](http://docs.nvidia.com/cuda/cuda-math-api/group __CUDA__ MATH __INTRINSIC__ INT.html)

---

<div class="post-metadata">

**Author:** ![njuffa](https://developer.download.nvidia.com/images/forums/profile-default-devtalk-84.png) [@njuffa](https://forums.developer.nvidia.com/u/njuffa)\
**Post date:** [January 28, 2018, 7:58am UTC](https://forums.developer.nvidia.com/t/handling-large-integer-calculations-inside-kernel-without-overflow/57518/3 "2018-01-28T07:58:01Z")

</div>

> [@](#):
>
> printf(“%ld\n”, \*c);

The safer choice of format here would be “%lld” since ‘long’ is a 32-bit type on some platforms (notably Windows).