Build issues on ROCm with random_ltd extension

[Hobbes-Le-Chat]

I saw this reported before, I am trying to install deepspeed suing the git repo

DS_BUILD_UTILS=1 DS_BUILD_CPU_ADAGRAD=1 DS_BUILD_RANDOM_LTD=1 DS_BUILD_CPU_ADAM=1 DS_BUILD_FUSED_ADAM=1 DS_BUILD_FUSED_LAMB=1 DS_BUILD_CCL_COMM=1 python setup.py build_ext -j80 bdist_wheel

I am hitting this error:

In file included from csrc/random_ltd/token_sort.hip:9:
In file included from csrc/includes/custom_hip_layers.h:17:
In file included from csrc/includes/context_hip.h:17:
csrc/includes/gemm_test_hip.h:312:13: warning: ignoring return value of function declared with 'nodiscard' attribute [-Wunused-result]
            hipDeviceSynchronize();
            ^~~~~~~~~~~~~~~~~~~~
csrc/includes/gemm_test_hip.h:306:13: warning: ignoring return value of function declared with 'nodiscard' attribute [-Wunused-result]
            hipDeviceSynchronize();
            ^~~~~~~~~~~~~~~~~~~~
csrc/includes/gemm_test_hip.h:222:24: note: in instantiation of function template specialization 'StridedGemmTest<__half>::Run<(lambda at csrc/includes/gemm_test_hip.h:222:35)>' requested here
        int algo_bw1 = Run(loops, [=](int algo) {
                       ^
csrc/includes/context_hip.h:160:53: note: in instantiation of member function 'StridedGemmTest<__half>::TestAlgo' requested here
            _gemm_algos.push_back(test_attn_scores->TestAlgo(100));
                                                    ^
In file included from csrc/random_ltd/token_sort.hip:9:
In file included from csrc/includes/custom_hip_layers.h:17:
In file included from csrc/includes/context_hip.h:17:
csrc/includes/gemm_test_hip.h:312:13: warning: ignoring return value of function declared with 'nodiscard' attribute [-Wunused-result]
            hipDeviceSynchronize();
            ^~~~~~~~~~~~~~~~~~~~
csrc/includes/gemm_test_hip.h:306:13: warning: ignoring return value of function declared with 'nodiscard' attribute [-Wunused-result]
            hipDeviceSynchronize();
            ^~~~~~~~~~~~~~~~~~~~
csrc/includes/gemm_test_hip.h:256:24: note: in instantiation of function template specialization 'StridedGemmTest<__half>::Run<(lambda at csrc/includes/gemm_test_hip.h:256:35)>' requested here
        int algo_bw2 = Run(loops, [=](int algo) {
                       ^
csrc/includes/context_hip.h:160:53: note: in instantiation of member function 'StridedGemmTest<__half>::TestAlgo' requested here
            _gemm_algos.push_back(test_attn_scores->TestAlgo(100));
                                                    ^
In file included from csrc/random_ltd/token_sort.hip:9:
In file included from csrc/includes/custom_hip_layers.h:17:
In file included from csrc/includes/context_hip.h:17:
csrc/includes/gemm_test_hip.h:312:13: warning: ignoring return value of function declared with 'nodiscard' attribute [-Wunused-result]
            hipDeviceSynchronize();
            ^~~~~~~~~~~~~~~~~~~~
17 warnings and 2 errors generated when compiling for gfx1030.
error: command '/opt/rocm-5.7.0/bin/hipcc' failed with exit code 1

Some info about my torch:

$ pip show torch 
Name: torch
Version: 2.2.0.dev20231128+rocm5.7
Summary: Tensors and Dynamic neural networks in Python with strong GPU acceleration
Home-page: https://pytorch.org/
Author: PyTorch Team
Author-email: packages@pytorch.org
License: BSD-3
Location: /env/rocm_57_p39_V2/lib/python3.9/site-packages
Requires: filelock, fsspec, jinja2, networkx, pytorch-triton-rocm, sympy, typing-extensions
Required-by: torchaudio, torchvision

[loadams]

FYI - @rraminen and @jithunnair-amd.

@Hobbes-Le-Chat - could you let us know what version of DeepSpeed you are using/share your ds_report?

[jithunnair-amd]

@Hobbes-Le-Chat I don't think you actually captured the error snippet, all we see are warnings and then:

17 warnings and 2 errors generated when compiling for gfx1030.
error: command '/opt/rocm-5.7.0/bin/hipcc' failed with exit code 1

Can you please provide a snippet of the actual error in your log?

[Hobbes-Le-Chat]

FYI - @rraminen and @jithunnair-amd.

@Hobbes-Le-Chat - could you let us know what version of DeepSpeed you are using/share your ds_report?

after removing DS_BUILD_RANDOM_LTD=1

$ DS_BUILD_UTILS=1 DS_BUILD_CPU_ADAGRAD=1 DS_BUILD_CPU_ADAM=1 DS_BUILD_FUSED_ADAM=1 DS_BUILD_FUSED_LAMB=1 DS_BUILD_CCL_COMM=1 python setup.py build_ext -j80 bdist_wheel

$ ds_report 
[2023-11-29 18:25:08,359] [INFO] [real_accelerator.py:161:get_accelerator] Setting ds_accelerator to cuda (auto detect)
--------------------------------------------------
DeepSpeed C++/CUDA extension op report
--------------------------------------------------
NOTE: Ops not installed will be just-in-time (JIT) compiled at
      runtime if needed. Op compatibility means that your system
      meet the required dependencies to JIT install the op.
--------------------------------------------------
JIT compiled ops requires ninja
ninja .................. [OKAY]
--------------------------------------------------
op name ................ installed .. compatible
--------------------------------------------------
 [WARNING]  async_io requires the dev libaio .so object and headers but these were not found.
 [WARNING]  If libaio is already installed (perhaps from source), try setting the CFLAGS and LDFLAGS environment variables to where it can be found.
async_io ............... [NO] ....... [NO]
fused_adam ............. [YES] ...... [OKAY]
cpu_adam ............... [YES] ...... [OKAY]
cpu_adagrad ............ [YES] ...... [OKAY]
cpu_lion ............... [NO] ....... [OKAY]
 [WARNING]  Please specify the CUTLASS repo directory as environment variable $CUTLASS_PATH
evoformer_attn ......... [NO] ....... [NO]
fused_lamb ............. [YES] ...... [OKAY]
fused_lion ............. [NO] ....... [OKAY]
inference_core_ops ..... [NO] ....... [OKAY]
cutlass_ops ............ [NO] ....... [OKAY]
quantizer .............. [NO] ....... [OKAY]
ragged_device_ops ...... [NO] ....... [OKAY]
ragged_ops ............. [NO] ....... [OKAY]
random_ltd ............. [NO] ....... [OKAY]
 [WARNING]  sparse_attn is not compatible with ROCM
sparse_attn ............ [NO] ....... [NO]
spatial_inference ...... [NO] ....... [OKAY]
transformer ............ [NO] ....... [OKAY]
stochastic_transformer . [NO] ....... [OKAY]
transformer_inference .. [NO] ....... [OKAY]
--------------------------------------------------
DeepSpeed general environment info:
torch install path ............... ['/rocm_57_p39_V2/lib/python3.9/site-packages/torch']
torch version .................... 2.2.0.dev20231128+rocm5.7
deepspeed install path ........... ['/rocm_57_p39_V2/lib/python3.9/site-packages/deepspeed']
deepspeed info ................... 0.12.4+8640b8e5, 8640b8e5, master
torch cuda version ............... None
torch hip version ................ 5.7.31921-d1770ee1b
nvcc version ..................... None
deepspeed wheel compiled w. ...... torch 2.2, hip 5.7
shared memory (/dev/shm) size .... 251.35 GB

[Hobbes-Le-Chat]

@Hobbes-Le-Chat I don't think you actually captured the error snippet, all we see are warnings and then:

17 warnings and 2 errors generated when compiling for gfx1030.
error: command '/opt/rocm-5.7.0/bin/hipcc' failed with exit code 1

Can you please provide a snippet of the actual error in your log?

Where can I find the log file?

[loadams]

@Hobbes-Le-Chat - does this mean that you do not see the error when removing the DS_BUILD_RANDOM_LTD=1?

And if you can just copy/paste the whole output from your console, that should have the error.

[Hobbes-Le-Chat]

@Hobbes-Le-Chat - does this mean that you do not see the error when removing the DS_BUILD_RANDOM_LTD=1?

And if you can just copy/paste the whole output from your console, that should have the error.

That is correct, if I remove S_BUILD_RANDOM_LTD=1 I don't see the error and I can install deepspeed.
@jithunnair-amd Is this the info you need?
deepspeedoutputerror.txt

[loadams]

@Hobbes-Le-Chat -

That's better, @jithunnair-amd these look to be the errors:

/opt/rocm-5.7.0/include/hip/amd_detail/amd_hip_cooperative_groups.h:638:3: error: static assertion failed due to requirement 'integral_constant<bool, false>::value': Tile size is either not a power of 2 or greater than the wavefront size

/opt/rocm-5.7.0/include/hip/amd_detail/amd_hip_cooperative_groups.h:723:18: error: no member named 'sync' in 'cooperative_groups::thread_block_tile_base<64>'

I'm not sure if we've tested everything on our side with ROCm 5.7 or not yet though.

[rraminen]

This is a known issue that we are looking into and as a work around comment the asserts in /opt/rocm-5.7.0/include/hip/amd_detail/amd_hip_cooperative_groups.h like below:

637 template <unsigned int size> class thread_block_tile_base : public tile_base<size> {
638   //static_assert(is_valid_tile_size<size>::value,
639   //              "Tile size is either not a power of 2 or greater than the wavefront size");
640   using tile_base<size>::numThreads;
 
836 __CG_QUALIFIER__ thread_block_tile<size, ParentCGTy> tiled_partition(const ParentCGTy& g) {
837   //static_assert(is_valid_tile_size<size>::value,
838   //              "Tiled partition with size > wavefront size. Currently not supported ");
839   return impl::tiled_partition_internal<size, ParentCGTy>(g);
840 }

[jithunnair-amd]

@Hobbes-Le-Chat Thanks, the log file helped!

Btw, I think you should update the title of this issue to "Build issues on ROCm with random_ltd extension" or something, since I don't think HIP_PLATFORM_AMD change has anything to do with the issues you're seeing.

[Hobbes-Le-Chat]

/opt/rocm-5.7.0/include/hip/amd_detail/amd_hip_cooperative_groups.h

I am working on a computational facility platform, thus I am not allow to edit those type of file. Is there a different work around?