CUDA atomicAdd执行异常：缓存不足时计数溢出及内存错误求助

AI 小北

2026-8-9

问题分析与解决方案

问题背景

尝试通过预设初始缓存大小初始化numpy矩阵，期望每个CUDA线程最多执行一次atomicAdd操作，保证累计计数不超过初始缓存大小。但遇到异常：当初始缓存大小（500）小于线程数（1024）时，累计计数出现异常极大值（1140850688），还触发内存错误。

第一种实现方案

代码

import os

_path = r"C:\Program Files\Microsoft Visual Studio\2022\Community\VC\Tools\MSVC\14.33.31629\bin\Hostx64\x64"

if os.system("cl.exe"):
    os.environ['PATH'] += ';' + _path
if os.system("cl.exe"):
    raise RuntimeError("cl.exe still not found, path probably incorrect")

import pycuda.driver as cuda
import pycuda.autoinit
from pycuda.compiler import SourceModule
import pandas as pd
import numpy as np

RESULT_COLUMN_COUNT = 4

InitialResultCacheSize = 10000
# InitialResultCacheSize = 500

number_matrix = np.zeros(InitialResultCacheSize * RESULT_COLUMN_COUNT)
number_matrix = number_matrix.astype(np.float32)

number_matrix_gpu = cuda.mem_alloc(number_matrix.nbytes)
cuda.memcpy_htod(number_matrix_gpu, number_matrix)

result_count = np.int32(0)
result_count_gpu = cuda.mem_alloc(result_count.nbytes)
cuda.memcpy_htod(result_count_gpu, result_count)

mod = SourceModule("""
  #include <cstdlib>

    __global__ void test_cuda_utilisation(int InitialResultCacheSize, int RESULT_COLUMN_COUNT, float *number_matrix, int *result_count)
  {
     int result_index, result_index_offset;
     
     if (result_count[0] < InitialResultCacheSize - 1) {
       result_index = atomicAdd(result_count,1);
       result_index_offset = result_index * RESULT_COLUMN_COUNT;
       number_matrix[result_index_offset + 0] = result_index;
       number_matrix[result_index_offset + 1] = result_count[0];
       number_matrix[result_index_offset + 2] = InitialResultCacheSize;
       number_matrix[result_index_offset + 3] = RESULT_COLUMN_COUNT;
     }
  }
  """)
      
func = mod.get_function("test_cuda_utilisation")
func(np.int32(InitialResultCacheSize), np.int32(RESULT_COLUMN_COUNT), number_matrix_gpu, result_count_gpu, block=(4,16,16))

result_count_out = np.empty_like(result_count)
cuda.memcpy_dtoh(result_count_out, result_count_gpu)
print('result_count_out = ' + str(result_count_out) + ' and InitialResultCacheSize is ' + str(InitialResultCacheSize))

number_matrix_out = np.empty((result_count_out, RESULT_COLUMN_COUNT), dtype=np.float32)
cuda.memcpy_dtoh(number_matrix_out, number_matrix_gpu)

print('number_matrix_out is with len ' + str(len(number_matrix_out)) + ' x ' + str(len(number_matrix_out[0])))
print(number_matrix_out)

运行结果

InitialResultCacheSize=10000时

'cl.exe' is not recognized as an internal or external command,
operable program or batch file.
Microsoft (R) C/C++ Optimizing Compiler Version 19.33.31629 for x64
Copyright (C) Microsoft Corporation.  All rights reserved.

result_count_out = 1024 and InitialResultCacheSize is 10000
number_matrix_out is with len 1024 x 4
[[0.000e+00 1.024e+03 1.000e+04 4.000e+00]
 [1.000e+00 1.024e+03 1.000e+04 4.000e+00]
 [2.000e+00 1.024e+03 1.000e+04 4.000e+00]
 ...
 [1.021e+03 1.024e+03 1.000e+04 4.000e+00]
 [1.022e+03 1.024e+03 1.000e+04 4.000e+00]
 [1.023e+03 1.024e+03 1.000e+04 4.000e+00]]

InitialResultCacheSize=500时

'cl.exe' is not recognized as an internal or external command,
operable program or batch file.
Microsoft (R) C/C++ Optimizing Compiler Version 19.33.31629 for x64
Copyright (C) Microsoft Corporation.  All rights reserved.

result_count_out = 1140850688 and InitialResultCacheSize is 500
Traceback (most recent call last):

  File C:\PythonProjects\TradeAnalysis\Test\TestCUDAUtilisation.py:67 in <module>
    cuda.memcpy_dtoh(number_matrix_out, number_matrix_gpu)

LogicError: cuMemcpyDtoH failed: invalid argument

第二种实现方案

代码

import os

# _path = r"D:\Program Files (x86)\Microsoft Visual Studio\2019\Community\VC\Tools\MSVC\14.28.29910\bin\Hostx64\x64"
_path = r"C:\Program Files\Microsoft Visual Studio\2022\Community\VC\Tools\MSVC\14.33.31629\bin\Hostx64\x64"
# _path = r"C:\Program Files\Microsoft Visual Studio\2022\Community\VC\Tools\MSVC\14.33.31629\bin\Hostx64"
# _path = r"C:\Program Files\Microsoft Visual Studio\2022\Community\VC\Tools\MSVC\14.33.31629\bin\Hostx64\x64\"

if os.system("cl.exe"):
    os.environ['PATH'] += ';' + _path
if os.system("cl.exe"):
    raise RuntimeError("cl.exe still not found, path probably incorrect")

import pycuda.driver as cuda
import pycuda.autoinit
from pycuda.compiler import SourceModule
import pandas as pd
import numpy as np

RESULT_COLUMN_COUNT = 4

# InitialResultCacheSize = 10000
InitialResultCacheSize = 500


number_matrix = np.zeros(InitialResultCacheSize * RESULT_COLUMN_COUNT)
number_matrix = number_matrix.astype(np.float32)

number_matrix_gpu = cuda.mem_alloc(number_matrix.nbytes)
cuda.memcpy_htod(number_matrix_gpu, number_matrix)

result_count = 0

mod = SourceModule("""
  #include <cstdlib>

    __global__ void test_cuda_utilisation(int InitialResultCacheSize, int RESULT_COLUMN_COUNT, int result_count, float *number_matrix)
  {
     int result_index, result_index_offset;
     
     result_index = atomicAdd(&result_count,1);
     if (result_index < InitialResultCacheSize - 1) {
       result_index_offset = result_index * RESULT_COLUMN_COUNT;
       number_matrix[result_index_offset + 0] = result_index;
       number_matrix[result_index_offset + 1] = result_count;
       number_matrix[result_index_offset + 2] = InitialResultCacheSize;
       number_matrix[result_index_offset + 3] = RESULT_COLUMN_COUNT;
     }
  }
  """)
      
func = mod.get_function("test_cuda_utilisation")
func(np.int32(InitialResultCacheSize), np.int32(RESULT_COLUMN_COUNT), np.int32(result_count), number_matrix_gpu, block=(4,16,16))

print('result_count = ' + str(result_count) + ' and InitialResultCacheSize = ' + str(InitialResultCacheSize))

运行结果

result_count = 0 and InitialResultCacheSize = 500

问题原因

第一种实现的竞态条件：判断result_count[0] < InitialResultCacheSize -1与执行atomicAdd并非原子操作。当多个线程同时通过判断时，会导致result_count超过缓存上限，甚至越界写入内存，破坏内存结构，最终使result_count出现乱码值，后续内存拷贝因目标数组大小错误触发异常。
第二种实现的参数传递错误：result_count被作为值传递给核函数，每个线程拿到的是本地副本，atomicAdd操作的是线程本地变量，完全不会修改主机端的result_count，因此主机端值始终为0。

正确实现方案

修正后的完整代码

import os

_path = r"C:\Program Files\Microsoft Visual Studio\2022\Community\VC\Tools\MSVC\14.33.31629\bin\Hostx64\x64"

if os.system("cl.exe"):
    os.environ['PATH'] += ';' + _path
if os.system("cl.exe"):
    raise RuntimeError("cl.exe still not found, path probably incorrect")

import pycuda.driver as cuda
import pycuda.autoinit
from pycuda.compiler import SourceModule
import numpy as np

RESULT_COLUMN_COUNT = 4
InitialResultCacheSize = 500

# 初始化GPU矩阵
number_matrix = np.zeros(InitialResultCacheSize * RESULT_COLUMN_COUNT, dtype=np.float32)
number_matrix_gpu = cuda.mem_alloc(number_matrix.nbytes)
cuda.memcpy_htod(number_matrix_gpu, number_matrix)

# 初始化计数变量（GPU端）
result_count = np.int32(0)
result_count_gpu = cuda.mem_alloc(result_count.nbytes)
cuda.memcpy_htod(result_count_gpu, result_count)

mod = SourceModule("""
    __global__ void test_cuda_utilisation(int InitialResultCacheSize, int RESULT_COLUMN_COUNT, float *number_matrix, int *result_count)
    {
        int result_index;
        // 先执行原子递增获取当前索引，再判断是否在缓存范围内
        result_index = atomicAdd(result_count, 1);
        if (result_index < InitialResultCacheSize) {
            int result_index_offset = result_index * RESULT_COLUMN_COUNT;
            number_matrix[result_index_offset + 0] = result_index;
            number_matrix[result_index_offset + 1] = result_count[0];
            number_matrix[result_index_offset + 2] = InitialResultCacheSize;
            number_matrix[result_index_offset + 3] = RESULT_COLUMN_COUNT;
        } else {
            // 超过上限时，将计数回退，保证最终计数不超过缓存大小
            atomicSub(result_count, 1);
        }
    }
""")

func = mod.get_function("test_cuda_utilisation")
# 启动1024个线程（4*16*16）
func(np.int32(InitialResultCacheSize), np.int32(RESULT_COLUMN_COUNT), number_matrix_gpu, result_count_gpu, block=(4,16,16))

# 拷贝结果回主机
result_count_out = np.empty_like(result_count)
cuda.memcpy_dtoh(result_count_out, result_count_gpu)
print(f'result_count_out = {result_count_out} and InitialResultCacheSize is {InitialResultCacheSize}')

# 仅拷贝有效数据部分
number_matrix_out = np.empty((result_count_out, RESULT_COLUMN_COUNT), dtype=np.float32)
cuda.memcpy_dtoh(number_matrix_out, number_matrix_gpu)

print(f'number_matrix_out is with len {len(number_matrix_out)} x {len(number_matrix_out[0])}')
print(number_matrix_out)

核心逻辑说明

先执行atomicAdd获取当前索引，再判断索引是否小于缓存大小：避免多线程竞态，确保每个线程仅执行一次原子操作。
若索引超过上限，用atomicSub将计数回退，保证最终result_count不超过初始缓存大小。
主机端拷贝结果时，使用正确的result_count_out大小初始化数组，避免内存拷贝错误。

内容的提问来源于stack exchange，提问作者Henry

CUDA atomicAdd执行异常：缓存不足时计数溢出及内存错误求助

问题背景 #

第一种实现方案 #

代码 #

运行结果 #

InitialResultCacheSize=10000时

InitialResultCacheSize=500时

第二种实现方案 #

代码 #

运行结果 #

问题原因 #

正确实现方案 #

修正后的完整代码 #

核心逻辑说明 #

问题背景

第一种实现方案

代码

运行结果

第二种实现方案

代码

运行结果

问题原因

正确实现方案

修正后的完整代码

核心逻辑说明