__constant__ で宣言した変数はホスト側で参照できない【CUDA】

CUDA

試した環境

本題

__constant__ で宣言した変数はデバイス側の変数ですが、ホスト側で使用するグローバル変数のような初期化の書き方ができます。 だからと言ってホスト側で直接参照できるわけではありません。
下記のコードでは__constant__ で宣言した変数Nをデバイス側ホスト側両方で参照しようとしています。

#include "cuda_runtime.h"
#include "device_launch_parameters.h"
#include <stdio.h>

__constant__ int N = 100;

__global__ void addConstant(int* dst, const int* src) {
    int i = threadIdx.x;
    dst[i] = src[i] + N;
}

int main()
{
    const int array_size = 5;
    const int src[array_size] = { 1, 2, 3, 4, 5 };
    int dst[array_size] = { 0 };

    int* dev_src = nullptr;
    int* dev_dst = nullptr;

    cudaMalloc((void**)&dev_src, array_size * sizeof(int));
    cudaMalloc((void**)&dev_dst, array_size * sizeof(int));

    cudaMemcpy(dev_src, src, array_size * sizeof(int), cudaMemcpyHostToDevice);

    addConstant <<<1, array_size >>> (dev_dst, dev_src);

    cudaMemcpy(dst, dev_dst, array_size * sizeof(int), cudaMemcpyDeviceToHost);

    printf("N = %d\n", N);
    printf("{1,2,3,4,5} + 100 = {%d,%d,%d,%d,%d}\n",
        dst[0], dst[1], dst[2], dst[3], dst[4]);

    cudaFree(dev_dst);
    cudaFree(dev_src);

    return 0;
}

上記ソースコードを実行した結果は以下の通りです。

N = 0
{1,2,3,4,5} + 100 = {101,102,103,104,105}

__constant__ で宣言した変数はデバイス側の変数なのでホスト側で参照に失敗します。 当たり前といえば当たり前なのですが、私はミスしてしまいました。