ROCM: Garbadge output

[Jipok]

GPTQ models works with exllama v1.

python test_inference.py -m ~/models/Synthia-13B-exl2 -p "Once upon a time,"
Successfully preprocessed all matching files.
 -- Model: /home/llama/models/Synthia-13B-exl2
 -- Options: ['rope_scale 1.0', 'rope_alpha 1.0']
 -- Loading model...
 -- Loading tokenizer...
 -- Warmup...
 -- Generating (greedy sampling)...

Once upon a time,ttt...............................................................................................................tttttttttttttt

Prompt processed in 0.10 seconds, 5 tokens, 51.99 tokens/second
Response generated in 3.96 seconds, 128 tokens, 32.29 tokens/second

$ python examples/inference.py
Successfully preprocessed all matching files.
Loading model: /home/llama/models/Synthia-13B-GPTQ/
Our story begins in the Scottish town of Auchtermuchty, where onceu at on/'s
m .'. p the. .tth from and and at f. bet1 hn
  : a4. [[t and in thet cd'
 research (Ft-t and e
 \({\f 701 346
s w56782 91,  ,·	 The08 " 710 and...6 1501020s	29
  

 @a70'27,[
 // 052
 ¡204; The
 %
4 this
 {5 it is just the s by some .

Response generated in 3.94 seconds, 150 tokens, 38.09 tokens/second

$ python examples/inference.py
Successfully preprocessed all matching files.
Loading model: /home/llama/models/Synthia-13B-exl2/
Our story begins in the Scottish town of Auchtermuchty, where onceo andt\\una
2​t andd​At t.th[t'ms
<,-d... , and03.0.	- ./,:
|m ont1. t605 thet7.th1  fy s to repv ag

....    The (p8628th.{{ 2l5-e.Zygt1t94hs0m.　
 | 57- f-n3, [[.[^-667. t8 and*1
Zyg7. | 3675, [[rF0th

Response generated in 5.25 seconds, 150 tokens, 28.59 tokens/second

GPU: AMD Instinct MI50
Name in OS: AMD ATI Radeon VII
Arch: gfx906

rocminfo

ROCk module is loaded
=====================
HSA System Attributes
=====================
Runtime Version:         1.1
System Timestamp Freq.:  1000.000000MHz
Sig. Max Wait Duration:  18446744073709551615 (0xFFFFFFFFFFFFFFFF) (timestamp count)
Machine Model:           LARGE
System Endianness:       LITTLE
...
*******
Agent 2
*******
  Name:                    gfx906
  Uuid:                    GPU-6f9a60e1732c7315
  Marketing Name:          AMD Radeon VII
  Vendor Name:             AMD
  Feature:                 KERNEL_DISPATCH
  Profile:                 BASE_PROFILE
  Float Round Mode:        NEAR
  Max Queue Number:        128(0x80)
  Queue Min Size:          64(0x40)
  Queue Max Size:          131072(0x20000)
  Queue Type:              MULTI
  Node:                    1
  Device Type:             GPU
  Cache Info:
    L1:                      16(0x10) KB
    L2:                      8192(0x2000) KB
  Chip ID:                 26287(0x66af)
  ASIC Revision:           1(0x1)
  Cacheline Size:          64(0x40)
  Max Clock Freq. (MHz):   1801
  BDFID:                   1280
  Internal Node ID:        1
  Compute Unit:            60
  SIMDs per CU:            4
  Shader Engines:          4
  Shader Arrs. per Eng.:   1
  WatchPts on Addr. Ranges:4
  Features:                KERNEL_DISPATCH
  Fast F16 Operation:      TRUE
  Wavefront Size:          64(0x40)
  Workgroup Max Size:      1024(0x400)
  Workgroup Max Size per Dimension:
    x                        1024(0x400)
    y                        1024(0x400)
    z                        1024(0x400)
  Max Waves Per CU:        40(0x28)
  Max Work-item Per CU:    2560(0xa00)
  Grid Max Size:           4294967295(0xffffffff)
  Grid Max Size per Dimension:
    x                        4294967295(0xffffffff)
    y                        4294967295(0xffffffff)
    z                        4294967295(0xffffffff)
  Max fbarriers/Workgrp:   32
  Pool Info:
    Pool 1
      Segment:                 GLOBAL; FLAGS: COARSE GRAINED
      Size:                    16760832(0xffc000) KB
      Allocatable:             TRUE
      Alloc Granule:           4KB
      Alloc Alignment:         4KB
      Accessible by all:       FALSE
    Pool 2
      Segment:                 GROUP
      Size:                    64(0x40) KB
      Allocatable:             FALSE
      Alloc Granule:           0KB
      Alloc Alignment:         0KB
      Accessible by all:       FALSE
  ISA Info:
    ISA 1
      Name:                    amdgcn-amd-amdhsa--gfx906:sramecc+:xnack-
      Machine Models:          HSA_MACHINE_MODEL_LARGE
      Profiles:                HSA_PROFILE_BASE
      Default Rounding Mode:   NEAR
      Default Rounding Mode:   NEAR
      Fast f16:                TRUE
      Workgroup Max Size:      1024(0x400)
      Workgroup Max Size per Dimension:
        x                        1024(0x400)
        y                        1024(0x400)
        z                        1024(0x400)
      Grid Max Size:           4294967295(0xffffffff)
      Grid Max Size per Dimension:
        x                        4294967295(0xffffffff)
        y                        4294967295(0xffffffff)
        z                        4294967295(0xffffffff)
      FBarrier Max Size:       32
*** Done ***

pytorch-lightning         1.9.4
pytorch-triton-rocm       2.1.0+34f8189eae
torch                     2.2.0.dev20230912+rocm5.6
torchaudio                2.2.0.dev20230912+rocm5.6
torchdiffeq               0.2.3
torchmetrics              1.1.2
torchsde                  0.2.5
torchvision               0.17.0.dev20230912+rocm5.6

[Jipok]

Slightly better result:

$ python examples/streaming.py
Successfully preprocessed all matching files.
Loading model: /home/llama/models/Marcoroni-7b-GPTQ/
Our story begins in the Scottish town of Auchtermuchty, where once.շ, this is a. I will be able to find it.A, as an alO2, the world and 107898, we’ll. 33.0460616,023507 12
J 394
O53190540.N663139
9.H4-T, A537475
0553,9
44720964-1749761 1991, 186003599503034209295:13,199 (2 5)73007130 4
014096033446
 5040 6.P181519B-4
79.W.4066
472732. 538145895345 1143
4371.4588

Prompt processed in 0.02 seconds, 15 tokens, 905.85 tokens/second
Response generated in 4.10 seconds, 250 tokens, 61.04 tokens/second

[ardfork]

Can you provide the exact model used? It works correctly on my machine:

python test_inference.py -m models/Llama-2-13B-chat-GPTQ -p "Once upon a time,"
Successfully preprocessed all matching files.
 -- Model: models/Llama-2-13B-chat-GPTQ
 -- Options: ['rope_scale 1.0', 'rope_alpha 1.0']
 -- Loading model...
 -- Loading tokenizer...
 -- Warmup...
 -- Generating (greedy sampling)...

Once upon a time, there was a young woman named Sophia who lived in a small village nestled in the rolling hills of Tuscany. Sophia was a beautiful and kind-hearted person, loved by all who knew her. She spent her days tending to her family's vineyard, and at night, she would sit by the fire and dream of adventure.

One evening, as she was gazing into the flames, a strange and handsome stranger appeared before her. He introduced himself as Leo, a traveler from a far-off land who had been wandering the countryside for many years

Prompt processed in 0.06 seconds, 5 tokens, 90.38 tokens/second
Response generated in 3.55 seconds, 128 tokens, 36.07 tokens/second

[Jipok]

@ardfork rocm version? GPU?

6_0-bpw-h6 https://huggingface.co/latimar/Synthia-13B-exl2
gptq-4bit-32g-actorder_True https://huggingface.co/TheBloke/Synthia-13B-GPTQ
gptq-4bit-32g-actorder_True https://huggingface.co/TheBloke/Marcoroni-7b-GPTQ

[ardfork]

I have ROCm 5.6.0 on my host, I have the same nightly PyTorch version as you. Using a 6700 xt with gfx1030 ISA.

I only tested Marcoroni-7b-GPTQ, similar to your last test:

python examples/streaming.py
Successfully preprocessed all matching files.
Loading model: models/Marcoroni-7b-GPTQ
Our story begins in the Scottish town of Auchtermuchty, where once upon a time there lived an old lady named Jeanie.
One day, as she was walking through her garden, she noticed that all the flowers were withering and dying. She knew it must be because they hadn't been watered properly. So, she decided to fetch some water from the well nearby.
As she walked towards the well, she heard a strange noise coming from inside it. Curious, she peeked into the well and saw a beautiful golden fish swimming around. The golden fish asked if she had any spare water for him, since he was so thirsty.
Jeanie was amazed by this extraordinary creature! She agreed to give him some water, but only if he promised to bring her good luck. The golden fish nodded his head eagerly, promising to do just that.
The next morning, when everyone woke up in Auchtermuchty, they discovered that their lives had changed for the better. The crops grew much faster than usual, and people became healthier and happier. It seemed like the golden fish kept its promise, bringing them all good fortune.
In return, the grateful villagers built a special pond near Jeanie's house

Prompt processed in 0.01 seconds, 15 tokens, 1003.77 tokens/second
Response generated in 4.63 seconds, 250 tokens, 53.95 tokens/second

So, it seems that either something is wrong when running on gfx906 or with your ROCm install.

Edit: I also verified that it works correctly with docker.io/rocm/dev-ubuntu-22.04:5.6.1-complete and latest ROCm nightly torch. Maybe try with a container to isolate potential ROCm installation problem.

[Jipok]

with your ROCm install.

GPTQ works on exllama v1. Transformers version works too.

[ardfork]

Are you using the same venv, particularly the same pytorch version?

I also have heard that sometime gfx906 can have some problem with some optimization level. Can you try replacing -O3 by -O2, -O1 and -O0 in exllamav2/ext.py? I think you only need to replace the one in extra_cuda_cflags but replace both just to be sure.

[fgdfgfthgr-fox]

Replacing -03 to -02 or -01 doesn't have any affect, still output garbadge. -00 cause memory access fault (reason: page not present or supervisor privilege) when warmup.

[SinanAkkoyun]

How did you compile exllamav2 to work with ROCm? Or did you just install these:

pytorch-lightning 1.9.4
pytorch-triton-rocm 2.1.0+34f8189eae
torch 2.2.0.dev20230912+rocm5.6
torchaudio 2.2.0.dev20230912+rocm5.6
torchdiffeq 0.2.3
torchmetrics 1.1.2
torchsde 0.2.5
torchvision 0.17.0.dev20230912+rocm5.6

In advance before installing exllamav2?

[Jipok]

Are you using the same venv, particularly the same pytorch version?

No venv. pytorch system-wide. Other pkgs installed for user.

Can you try replacing -O3 by -O2, -O1 and -O0 in exllamav2/ext.py?

No noticeable effect with -O1 and -O0. Anyway my GPU works with stable diffusion and exllama v1.

How did you compile exllamav2 to work with ROCm? In advance before installing exllamav2?

Yes. Just pip3 install --force --pre torch torchvision torchaudio --index-url https://download.pytorch.org/whl/nightly/rocm5.6

[ardfork]

Since you are not the only one with a gfx906 that have this problem. There is probably something wrong with ROCm and that GPU. Either some emitted code that is incorrect or a builtin function that is broken.

I'm curious if it has something to do with one of the half2 functions. Can you try this exllama v1 PR: turboderp/exllama#146? If it produce gibberish we can at least isolate the problem to one of the half2 function.

[Jipok]

Can you try this exllama v1 PR: turboderp/exllama#146?

Same error in AutoGPTQ/AutoGPTQ#340:

[7/10] /opt/rocm/bin/hipcc  -DWITH_HIP -DTORCH_EXTENSION_NAME=exllama_ext -DTORCH_API_INCLUDE_EXTENSION_H -DPYBIND11_COMPILER_TYPE=\"_gcc\" -DPYBIND11_STDLIB=\"_libstdcpp\" -DPYBIND11_BUILD_ABI=\"_cxxabi1011\" -I/home/llama/exllama-half2/exllama_ext -isystem /usr/lib/python3.11/site-packages/torch/include -isystem /usr/lib/python3.11/site-packages/torch/include/torch/csrc/api/include -isystem /usr/lib/python3.11/site-packages/torch/include/TH -isystem /usr/lib/python3.11/site-packages/torch/include/THC -isystem /usr/lib/python3.11/site-packages/torch/include/THH -isystem /opt/rocm/include -isystem /usr/include/python3.11 -D_GLIBCXX_USE_CXX11_ABI=0 -fPIC -std=c++17 -O3 -fPIC -D__HIP_PLATFORM_HCC__=1 -DUSE_ROCM=1 -DCUDA_HAS_FP16=1 -D__HIP_NO_HALF_OPERATORS__=1 -D__HIP_NO_HALF_CONVERSIONS__=1 -lineinfo -O3 --offload-arch=gfx906 -fno-gpu-rdc -c /home/llama/exllama-half2/exllama_ext/hip_func/q4_matmul.hip -o q4_matmul.cuda.o
FAILED: q4_matmul.cuda.o
/opt/rocm/bin/hipcc  -DWITH_HIP -DTORCH_EXTENSION_NAME=exllama_ext -DTORCH_API_INCLUDE_EXTENSION_H -DPYBIND11_COMPILER_TYPE=\"_gcc\" -DPYBIND11_STDLIB=\"_libstdcpp\" -DPYBIND11_BUILD_ABI=\"_cxxabi1011\" -I/home/llama/exllama-half2/exllama_ext -isystem /usr/lib/python3.11/site-packages/torch/include -isystem /usr/lib/python3.11/site-packages/torch/include/torch/csrc/api/include -isystem /usr/lib/python3.11/site-packages/torch/include/TH -isystem /usr/lib/python3.11/site-packages/torch/include/THC -isystem /usr/lib/python3.11/site-packages/torch/include/THH -isystem /opt/rocm/include -isystem /usr/include/python3.11 -D_GLIBCXX_USE_CXX11_ABI=0 -fPIC -std=c++17 -O3 -fPIC -D__HIP_PLATFORM_HCC__=1 -DUSE_ROCM=1 -DCUDA_HAS_FP16=1 -D__HIP_NO_HALF_OPERATORS__=1 -D__HIP_NO_HALF_CONVERSIONS__=1 -lineinfo -O3 --offload-arch=gfx906 -fno-gpu-rdc -c /home/llama/exllama-half2/exllama_ext/hip_func/q4_matmul.hip -o q4_matmul.cuda.o
clang-16: warning: -lineinfo: 'linker' input unused [-Wunused-command-line-argument]
In file included from /home/llama/exllama-half2/exllama_ext/hip_func/q4_matmul.hip:5:
/home/llama/exllama-half2/exllama_ext/hip_func/../util_hip.cuh:44:5: warning: ignoring return value of function declared with 'nodiscard' attribute [-Wunused-result]
    hipDeviceSynchronize();
    ^~~~~~~~~~~~~~~~~~~~
/home/llama/exllama-half2/exllama_ext/hip_func/../util_hip.cuh:58:5: warning: ignoring return value of function declared with 'nodiscard' attribute [-Wunused-result]
    hipDeviceSynchronize();
    ^~~~~~~~~~~~~~~~~~~~
/home/llama/exllama-half2/exllama_ext/hip_func/q4_matmul.hip:185:5: warning: ignoring return value of function declared with 'nodiscard' attribute [-Wunused-result]
    hipSetDevice(w->device);
    ^~~~~~~~~~~~ ~~~~~~~~~
/home/llama/exllama-half2/exllama_ext/hip_func/q4_matmul.hip:236:5: warning: ignoring return value of function declared with 'nodiscard' attribute [-Wunused-result]
    hipSetDevice(w->device);
    ^~~~~~~~~~~~ ~~~~~~~~~
/home/llama/exllama-half2/exllama_ext/hip_func/q4_matmul.hip:260:5: error: no matching function for call to 'hipblasHgemm'
    hipblasHgemm(handle, HIPBLAS_OP_N, HIPBLAS_OP_N, width, height, dim, &alpha, buffers->temp_dq, width, x_mapped, dim, &beta, out, width);
    ^~~~~~~~~~~~
/opt/rocm/include/hipblas/hipblas.h:10636:32: note: candidate function not viable: no known conversion from 'const half *' (aka 'const __half *') to 'const hipblasHalf *' (aka 'const unsigned short *') for 7th argument
HIPBLAS_EXPORT hipblasStatus_t hipblasHgemm(hipblasHandle_t    handle,
                               ^

I'm curious if it has something to do with one of the half2 functions.

Idk. I have installed AUTOMATIC1111/stable-diffusion-webui and wiki says:
*For many AMD GPUs, you must add --precision full --no-half or --upcast-sampling arguments to avoid NaN errors or crashing.

For me both --precision full --no-half and without args works. But without args i have ~2x less ram consumption:

Two runs

[ardfork]

Same error

What do you mean same error? It's the first time you are posting that one.

This error is because it is based on an older exllama version, it is broken on recent pytorch. Apply the PR on current exllama master or backport this patch turboderp/exllama@3dff8fe.

For me both --precision full --no-half and without args works. But without args i have ~2x less ram consumption:

Exllama always use half, but in exllama v1, it could use half2 and half, half2 were broken (without the PR I linked) so it always used half. Exllama v2 support only half2 (saw some code for half support but don't think it's complete).

[Jipok]

What do you mean same error? It's the first time you are posting that one.

Updated post. I got a little confused responding to different discussions on a similar topic at the same time.

[Jipok]

This error is because it is based on an older exllama version, it is broken on recent pytorch. Apply the PR on current exllama master or backport this patch turboderp/exllama@3dff8fe.

In commit: ...this is fixed in ROCm 5.6.
So i skipped merging/reabse. Just used turboderp/exllama#146 with extra_cflags = ["-O3 -DHIPBLAS_USE_HIP_HALF=1"] in cuda_ext.py, like recommended in AutoGPTQ/AutoGPTQ#340.

Can you try this exllama v1 PR: turboderp/exllama#146?

Works, no garbage: PYTORCH_ROCM_ARCH=gfx906 python example_chatbot.py -d ~/models/Synthia-13B-GPTQ -un "Jipok" -p prompt_chatbort.txt