type
status
date
slug
summary
tags
icon
password
在试图白嫖FuseWindowProcess时，发现window_partition的前向没有问题，但是window_merge的前向在推理过程中经常容易爆错，本篇博客用于解决一下这个问题。
出事的代码如下：
问题描述与定位
上面这串代码会在一些输入分辨率非方形的情况下（例如分辨率在256*128）在推理过程中爆错。
Debug一下
我们首先编译一下debug版本cuda然后进入cuda-gdb
GDB，启动！
按r运行之后出现这样一个问题。
bt一下：
这里有一个函数：__ldg，Load Global Device，他的作用是：将全局内存中的数据放入缓存中，从而提高读取速度并确保数据的一致性。在cuda文档中，有其详细的描述：
Data that is read-only for the entire lifetime of the kernel can also be cached in the read-only data cache described in the previous section by reading it using the __ldg() function (see Read-Only Data Cache Load Function). When the compiler detects that the read-only condition is satisfied for some data, it will use __ldg() to read it. The compiler might not always be able to detect that the read-only condition is satisfied for some data. Marking pointers used for loading such data with both the const and restrict qualifiers increases the likelihood that the compiler will detect the read-only condition.
一般来说__ldg不会有问题，input是输入，output是输出，我们先排除这俩的问题，比如尽量避免出现nan。我们在WindowProcessReverse调用前后都添加上assert：
运行之后报错，但是报错语句出现在reshape后：
这就很诡异了，这意味着WindowProcessReverse的输出是正常的，但是经过了reshape后出现了OOM问题。目前不清楚是reshape的问题还是x自己的问题，于是我多尝试了几种：
报错如下：
可以见到，对x进行任何操作，包括inplace操作，都会导致illegal memory access，也就是说x本身可能指向了一些非法内存，定位到下面这行语句：
解决问题
原来就是这个地方应该对宽nW进行处理，原来大佬有时候也会犯这种小错误。
散会！