# Cmake and and Heterogenious GPUs

**URL:** <https://forums.developer.nvidia.com/t/cmake-and-and-heterogenious-gpus/18793>\
**Category:** CUDA Programming and Performance\
**Created:** [September 14, 2010, 9:48pm UTC](https://forums.developer.nvidia.com/t/cmake-and-and-heterogenious-gpus/18793 "2010-09-14T21:48:33Z")\
**Posts on this page:** 13\
**Page:** 1

<div class="post-metadata">

**Author:** ![enemyben88](https://developer.download.nvidia.com/images/forums/profile-default-devtalk-84.png) [@enemyben88](https://forums.developer.nvidia.com/u/enemyben88)\
**Post date:** [September 14, 2010, 9:48pm UTC](https://forums.developer.nvidia.com/t/cmake-and-and-heterogenious-gpus/18793/1 "2010-09-14T21:48:33Z")

</div>

Hi all,

Just wondering if anyone has done any multi-GPU applications involving heterogeneous GPUs of different compute capabilities? Specifically, what I am trying to do is split my algorithm/data across a version 1.3 device and a 2.0 device, compiled with their respective device versions (-arch = sm\_13, -arch=sm\_20) using FindCUDA.cmake. Anyone done something like this?

Thanks!

---

<div class="post-metadata">

**Author:** ![zeus13i](https://developer.download.nvidia.com/images/forums/profile-default-devtalk-84.png) [@zeus13i](https://forums.developer.nvidia.com/u/zeus13i)\
**Post date:** [September 15, 2010, 6:32am UTC](https://forums.developer.nvidia.com/t/cmake-and-and-heterogenious-gpus/18793/2 "2010-09-15T06:32:18Z")

</div>

Hi Andy,

I haven’t looked into this too extensively, but hopefully this’ll help you get started:

In terms of compilation, I believe you have to make use of “CUDA\_NVCC\_FLAGS” in your CMakeLists.txt (from FindCUDA)in conjunction with the “-gencode” nvcc flag.

In terms of code, you can use the “ **CUDA\_ARCH** ” macro in your device code.

See 3.1.4 of Programming Guide 3.1

---

<div class="post-metadata">

**Author:** ![zeus13i](https://developer.download.nvidia.com/images/forums/profile-default-devtalk-84.png) [@zeus13i](https://forums.developer.nvidia.com/u/zeus13i)\
**Post date:** [September 15, 2010, 6:32am UTC](https://forums.developer.nvidia.com/t/cmake-and-and-heterogenious-gpus/18793/3 "2010-09-15T06:32:18Z")

</div>

Hi Andy,

I haven’t looked into this too extensively, but hopefully this’ll help you get started:

In terms of compilation, I believe you have to make use of “CUDA\_NVCC\_FLAGS” in your CMakeLists.txt (from FindCUDA)in conjunction with the “-gencode” nvcc flag.

In terms of code, you can use the “ **CUDA\_ARCH** ” macro in your device code.

See 3.1.4 of Programming Guide 3.1

---

<div class="post-metadata">

**Author:** ![Sarnath](https://sea2.discourse-cdn.com/nvidia/user_avatar/forums.developer.nvidia.com/sarnath/32/9742_2.png) [@Sarnath](https://forums.developer.nvidia.com/u/Sarnath)\
**Post date:** [September 15, 2010, 6:51am UTC](https://forums.developer.nvidia.com/t/cmake-and-and-heterogenious-gpus/18793/4 "2010-09-15T06:51:28Z")

</div>

My Dear Enemy,

Is it possible to tell NVCC to compile for multiple archs (like -arch 1.0,2.0,3.0 etc.)? Then, one can easily update FindCuda.Make

---

<div class="post-metadata">

**Author:** ![Sarnath](https://sea2.discourse-cdn.com/nvidia/user_avatar/forums.developer.nvidia.com/sarnath/32/9742_2.png) [@Sarnath](https://forums.developer.nvidia.com/u/Sarnath)\
**Post date:** [September 15, 2010, 6:51am UTC](https://forums.developer.nvidia.com/t/cmake-and-and-heterogenious-gpus/18793/5 "2010-09-15T06:51:28Z")

</div>

My Dear Enemy,

Is it possible to tell NVCC to compile for multiple archs (like -arch 1.0,2.0,3.0 etc.)? Then, one can easily update FindCuda.Make

---

<div class="post-metadata">

**Author:** ![MisterAnderson42](https://developer.download.nvidia.com/images/forums/profile-default-devtalk-84.png) [@MisterAnderson42](https://forums.developer.nvidia.com/u/MisterAnderson42)\
**Post date:** [September 15, 2010, 12:22pm UTC](https://forums.developer.nvidia.com/t/cmake-and-and-heterogenious-gpus/18793/6 "2010-09-15T12:22:25Z")

</div>

The easy way to do this is to do what the Fermi compatibility guide recommends and add the following command line arguments.

```auto
list(APPEND CUDA_NVCC_FLAGS "-gencode=arch=compute_11,code=sm_11")

list(APPEND CUDA_NVCC_FLAGS "-gencode=arch=compute_11,code=compute_11")

list(APPEND CUDA_NVCC_FLAGS "-gencode=arch=compute_20,code=sm_20")

list(APPEND CUDA_NVCC_FLAGS "-gencode=arch=compute_20,code=compute_20")

```

Where you include another set of lines for every system you want to compile for (I currently just use 11 and 20 because I don’t use any 13 features).

Then, in the code, you can use **CUDA\_ARCH** in device code to prevent compile errors

```auto
__global__ void some_kernel(...)

	{

	// do something that compiles on the lowest compute capability

	}

__global__ void some_kernel_sm20(...)

	{

	#if __CUDA_ARCH__ >= 200

	// do something that uses compute 2.0 features

	#endif

	}

```

And then make a runtime decision based on info from cudaDeviceGetProperties to determine which kernel to call.

Another option is to break your sm\_11 and sm\_20 kernels into separate files and pass the appropriate arch flags to CUDA\_COMPILE for that file (see the FindCUDA.cmake docs for details). This way, you don’t need to use the #ifs, but it does require a bit more cmake code and well thought out file organization to manage it. The other advantage is that you don’t have a compiled sm\_20 version of som\_kernel wasting disk space that will never be called.

---

<div class="post-metadata">

**Author:** ![MisterAnderson42](https://developer.download.nvidia.com/images/forums/profile-default-devtalk-84.png) [@MisterAnderson42](https://forums.developer.nvidia.com/u/MisterAnderson42)\
**Post date:** [September 15, 2010, 12:22pm UTC](https://forums.developer.nvidia.com/t/cmake-and-and-heterogenious-gpus/18793/7 "2010-09-15T12:22:25Z")

</div>

The easy way to do this is to do what the Fermi compatibility guide recommends and add the following command line arguments.

```auto
list(APPEND CUDA_NVCC_FLAGS "-gencode=arch=compute_11,code=sm_11")

list(APPEND CUDA_NVCC_FLAGS "-gencode=arch=compute_11,code=compute_11")

list(APPEND CUDA_NVCC_FLAGS "-gencode=arch=compute_20,code=sm_20")

list(APPEND CUDA_NVCC_FLAGS "-gencode=arch=compute_20,code=compute_20")

```

Where you include another set of lines for every system you want to compile for (I currently just use 11 and 20 because I don’t use any 13 features).

Then, in the code, you can use **CUDA\_ARCH** in device code to prevent compile errors

```auto
__global__ void some_kernel(...)

	{

	// do something that compiles on the lowest compute capability

	}

__global__ void some_kernel_sm20(...)

	{

	#if __CUDA_ARCH__ >= 200

	// do something that uses compute 2.0 features

	#endif

	}

```

And then make a runtime decision based on info from cudaDeviceGetProperties to determine which kernel to call.

Another option is to break your sm\_11 and sm\_20 kernels into separate files and pass the appropriate arch flags to CUDA\_COMPILE for that file (see the FindCUDA.cmake docs for details). This way, you don’t need to use the #ifs, but it does require a bit more cmake code and well thought out file organization to manage it. The other advantage is that you don’t have a compiled sm\_20 version of som\_kernel wasting disk space that will never be called.

---

<div class="post-metadata">

**Author:** ![enemyben88](https://developer.download.nvidia.com/images/forums/profile-default-devtalk-84.png) [@enemyben88](https://forums.developer.nvidia.com/u/enemyben88)\
**Post date:** [September 26, 2010, 11:46pm UTC](https://forums.developer.nvidia.com/t/cmake-and-and-heterogenious-gpus/18793/8 "2010-09-26T23:46:37Z")

</div>

Thanks all, it worked great!

---

<div class="post-metadata">

**Author:** ![enemyben88](https://developer.download.nvidia.com/images/forums/profile-default-devtalk-84.png) [@enemyben88](https://forums.developer.nvidia.com/u/enemyben88)\
**Post date:** [September 26, 2010, 11:46pm UTC](https://forums.developer.nvidia.com/t/cmake-and-and-heterogenious-gpus/18793/9 "2010-09-26T23:46:37Z")

</div>

Thanks all, it worked great!

---

<div class="post-metadata">

**Author:** ![enemyben88](https://developer.download.nvidia.com/images/forums/profile-default-devtalk-84.png) [@enemyben88](https://forums.developer.nvidia.com/u/enemyben88)\
**Post date:** [September 26, 2010, 11:46pm UTC](https://forums.developer.nvidia.com/t/cmake-and-and-heterogenious-gpus/18793/10 "2010-09-26T23:46:37Z")

</div>

Thanks all, it worked great!

---

<div class="post-metadata">

**Author:** ![E.D\_Riedijk](https://developer.download.nvidia.com/images/forums/profile-default-devtalk-84.png) [@E.D\_Riedijk](https://forums.developer.nvidia.com/u/E.D_Riedijk)\
**Post date:** [September 27, 2010, 4:15am UTC](https://forums.developer.nvidia.com/t/cmake-and-and-heterogenious-gpus/18793/11 "2010-09-27T04:15:02Z")

</div>

> [@](#):
>
> Where you include another set of lines for every system you want to compile for (I currently just use 11 and 20 because I don’t use any 13 features).

Just a quick question from someone who doesn’t need to support multiple targets (yet).

If you generate sm\_11 code, does it still get optimized for 1.2 & 1.3 hardware (double amount of registers)?

---

<div class="post-metadata">

**Author:** ![E.D\_Riedijk](https://developer.download.nvidia.com/images/forums/profile-default-devtalk-84.png) [@E.D\_Riedijk](https://forums.developer.nvidia.com/u/E.D_Riedijk)\
**Post date:** [September 27, 2010, 4:15am UTC](https://forums.developer.nvidia.com/t/cmake-and-and-heterogenious-gpus/18793/12 "2010-09-27T04:15:02Z")

</div>

> [@](#):
>
> Where you include another set of lines for every system you want to compile for (I currently just use 11 and 20 because I don’t use any 13 features).

Just a quick question from someone who doesn’t need to support multiple targets (yet).

If you generate sm\_11 code, does it still get optimized for 1.2 & 1.3 hardware (double amount of registers)?

---

<div class="post-metadata">

**Author:** ![E.D\_Riedijk](https://developer.download.nvidia.com/images/forums/profile-default-devtalk-84.png) [@E.D\_Riedijk](https://forums.developer.nvidia.com/u/E.D_Riedijk)\
**Post date:** [September 27, 2010, 4:15am UTC](https://forums.developer.nvidia.com/t/cmake-and-and-heterogenious-gpus/18793/13 "2010-09-27T04:15:02Z")

</div>

> [@](#):
>
> Where you include another set of lines for every system you want to compile for (I currently just use 11 and 20 because I don’t use any 13 features).

Just a quick question from someone who doesn’t need to support multiple targets (yet).

If you generate sm\_11 code, does it still get optimized for 1.2 & 1.3 hardware (double amount of registers)?