```cpp
#include
#include
#include
// 错误检查宏
#define CUDA_CHECK(call) \
do { \
cudaError_t err = call; \
if (err != cudaSuccess) { \
std::cerr << "CUDA error: " << cudaGetErrorString(err) \
<< " at " << __FILE__ << ":" << __LINE__ << std::endl; \
exit(1); \
} \
} while(0)
#define checkCudaKernel(...) \
__VA_ARGS__; \
do{cudaError_t cudaStatus = cudaPeekAtLastError(); \
if (cudaStatus != cudaSuccess){ \
printf("launch failed: %s\n", cudaGetErrorString(cudaStatus)); \
}} while(0);
// 内核:尝试访问越界内存
__global__ void out_of_bounds_kernel(int* data, size_t size) {
int idx = blockIdx.x * blockDim.x + threadIdx.x;
// 故意越界访问
int offset = 100000;
data[idx + offset] = 42; // 访问超出分配内存范围
}
// 内核:使用未初始化的指针
__global__ void uninitialized_pointer_kernel(int* data) {
int* bad_ptr = nullptr;
*bad_ptr = 42; // 解引用空指针
}
// 内核:使用已释放的内存
__global__ void use_freed_memory_kernel(int* data) {
data[0] = 42; // data可能已经被释放
}
int main() {
std::cout << "=== CUDA Memory Aperture Violation Tests ===" << std::endl;
// 测试1: 越界访问
std::cout << "\nTest 1: Out-of-bounds access" << std::endl;
size_t size = 10;
int* d_data;
CUDA_CHECK(cudaMalloc(&d_data, size * sizeof(int)));
// 启动kernel,访问超出范围的内存
int threads = 256;
int blocks = 1;
for (int i = 0; i<100000; ++i){
out_of_bounds_kernel<<>>(d_data, size);
CUDA_CHECK(cudaPeekAtLastError());
uninitialized_pointer_kernel<<>>(d_data);
CUDA_CHECK(cudaPeekAtLastError());
CUDA_CHECK(cudaDeviceSynchronize()); // 这里应该会报错
}
CUDA_CHECK(cudaFree(d_data));
use_freed_memory_kernel<<>>(d_data);
CUDA_CHECK(cudaPeekAtLastError());
std::cout << "\nTest 1: Out-of-bounds access success!!!" << std::endl;
return 0;
}
```