kernel benchmark

[Dujianhua1008]

Hello,guys

Thank you for your work on FlatQuant. I am currently testing the kernel performance and have encountered some issues:

Kernel Testing Errors:

While qattention.py runs successfully, other tests result in errors such as

 FlatQuant/third-party/cutlass/include/cutlass/integer_subbyte.h:96: cutlass::integer_subbyte<Bits, Signed>::integer_subbyte(int) [with int Bits = 4; __nv_bool Signed = true]: block: [45,5,0], thread: [30,7,0] Assertion `value < upper_bound` failed.
Assertion failed: Error Internal

Import Statement Clarification:
In the deploy module, there is an import statement: import deploy._CUDA.​
I am unable to locate the corresponding module or understand its implementation.​
Could you please provide guidance on resolving the kernel testing errors and clarify the purpose and implementation of the deploy._CUDA module?
Looking forward to your reply.

[han65487312]

It seems like the cutlass kernel accesses illegal memory. You may test the kernel using RTX3090 following our environment setup guidance. And you can provide more details about the environment, GPU version, Pip list, and CUDA version, etc.